Skip to content

Populating pointers of struct from cupy #211

Description

@neworderofjamie

Not really an issue but, I'm trying to populate a struct in constant memory containing (device) pointers in order to run my kernel through kernel tuner. In C the struct looks something like:

struct MergedNeuronUpdateGroup0
 {
    unsigned int* spkCnt;
    unsigned int* spk;
};
__device__ __constant__ MergedNeuronUpdateGroup0 d_mergedNeuronUpdateGroup0[8];

I can see how I would pack pointers into this with struct.pack("PP0l", spk_cnt, spk) and then convert to numpy as described in your example before passing via the cmem_args kwarg to tune_kernel. But, I've no idea how I would create (and initialise) these pointers from Python! Any pointers?

Thanks in advance for your help!

Activity

  1. benvanwerkhoven commented on Sep 29, 2023

    @benvanwerkhoven
    Collaborator

    I'm not sure how to do this from cupy, but with pycuda it should be easy. If you have PyCuda device allocation you can just cast it to an int to get the GPU memory address. From a cupy array you may be able to extract the pointer somehow, but I just had a look at cupy and DLpack, but I'm not sure how to extract the device pointer from those data types. If using PyCuda is an option for you, I think you could make it work.

  2. neworderofjamie commented on Oct 2, 2023

    @neworderofjamie
    Author

    I haven't used either PyCuda or cupy so happy to use whichever is easier - thanks for the suggestion

  3. locked and limited conversation to collaborators on Nov 8, 2023
  4. converted this issue into a discussion #227 on Nov 8, 2023
Sign up for free to join this conversation on GitHub. Already have an account? Sign in to comment

Metadata

Metadata

Assignees

No one assigned

    Labels

    No labels
    No labels

    Type

    No type

    Projects

    No projects

      Milestone

      No milestone

      Relationships

      None yet

      Development

      No branches or pull requests

      Issue actions