2d thrust::device_vector to kernel

Question

I have a data type defined as typedef InitialState float[12]. I have a vector containing several initial states defined by std::vector<InitialState> h_initials.

I made it a thrust device vector : thrust::device_vector<InitialState> d_initials = h_initials;

I would like to use this data is a user-defined kernel. However, I'm having problems casting it to a raw pointer. I've tried

float *p_initials = thrust::raw_pointer_cast(&d_initials[0]);

but thrust complains function returning array is not allowed.

Is there a way I can cast the device vector to a pointer that the kernel can use?

Answer 1

InitialState (float[12]) != float

InitialState *p_initials = thrust::raw_pointer_cast(d_initials.data());
float* p_floats = (float*)p_initials;

However, this is generally wrong to start with, because of the weird behavior below

typedef int int4[4];
void test(int4 a)
{
   std::cout << sizeof(a) << std::endl;
}

int main(int argc, char** argv)
{
  int4 f;
  std::cout << sizeof(f) << std::endl;//returns 16 (4*sizeof(int))
  test(f);//returns 8 (size of pointer, passes by reference)
}

Whats better is this:

struct InitialState{float value[12];}
std::vector<InitialState> h_initials;
thrust::device_vector<InitialState> d_initials = h_initials;
InitialState *p_initials = thrust::raw_pointer_cast(d_initials.data());
float* p_floats = (float*)p_initials;

And in cuda, you can use either InitialState* or float* as the argument (though SOA works better than AOS)