pass pointer to device memory to kernel per struct


on the host i allocate memory for the device using cudaMalloc() which allocates device memory, according to the docs.
I have a large struct with heaps of parameters for the kernel:

typedef struct {

int* dev_mem;


The kernel looks like this:

global void my_kernel(MY_STRUCT* s) {
int tid = …
int* ip = s->dev_mem + 12 * tid;

ip[0] = 0; // Warning on this line

When compiling i get a warning:
Warning: Cannot tell what pointer points to, assuming global memory space

The problem is that the memory is NOT shared.

How can i tell the compiler that this is memory of the device?

Or is there any other way i need to change the code to make it work?

Thanks for any hints,

First of all, are you sure that you want to pass a pointer to the struct to the kernel? This only works if you construct the struct on the GPU. You probably want to pass the struct by value rather than by reference.

As for the warning, it means that it doesn’t know whether this is global device memory or shared device memory. It assumed the safe assumption that it is global memory.