Passing pointers in a struct

I have a set of data, buffers and metadata, that I wish to send to a kernel stored in a struct such as


typedef struct {
  __global float *buffer1;
  __global float *buffer2;
  float value;
  uint numElements;
} ComplexData;

__kernel void process(__constant ComplexData *data) {
  ...
}

I tried to make a host side struct where the float pointers were replaced by cl_mems, but that did not work.


typedef struct {
  cl_mem buffer1;
  cl_mem buffer2;
  float value;
  uint numElements;
} ComplexData;

The application either crashed in clBuildProgram trying to read from address 0x10, or gave garbage when referencing members in the struct. Calling sizeof() on the two structs, one in host code and one in device code, gives different sizes.

What is the correct way to pass a struct like this, or is the only option to send the members one by one?

The

typedef struct {
cl_mem buffer1;
cl_mem buffer2;
float value;
uint numElements;
} ComplexData;

is not valid in OpenCL. You cannot declare memory objects (buffer and image) in a struct. These must be passed as arguments to a kernel.

Having said this, this should return in an error returned by clBuildProgram. clBuildProgram should not crash though.

I haven’t tried this with OpenCL, but I believe you just need to do some marshelling. Just allocate gpu memory for each of the pointers in your host struct and then copy this struct to the gpu.

cl_mems are objects that track an internal representation of the data. You can’t pass them in to the GPU as they aren’t pointers. (Technically they’re pointers to an object, but they are not pointers to the data.) The only way to access a cl_mem from within a kernel is to pass it in as an argument via setKernelArgs. Even if they were pointers, you’d have problems because the GPU and CPU address spaces are different, so they wouldn’t even point to the right place.