Pointer arithmetic in device memory

I was wondering if anyone had experience using pointer arithmetic with device memory.

What I’m trying to do, is something like this:

extern float* array1;

extern int array1Size;

extern float* array2;

extern int array2Size;

__host__ void someFn( void )


  float* d_array;

 cudaMalloc( (void**)&d_array, sizeof(float) * (array1Size + array2Size) );

  cudaMemcpy( d_array, array1, sizeof(float) * array1Size, cudaMemcpyHostToDevice );

  cudaMemcpy( d_array + (sizeof(float) * array1Size), array2, sizeof(float) * array2Size, cudaMemcpyHostToDevice );

 someKernel<<<gridSize, blockSize>>>( d_array );


I’ve had this code in my kernel for a while, and everything seemed to be working, but all of the sudden I get this error on the second cudaMemcpy call in debug emulation mode:

Cuda error in file: invalid device pointer.

Am I supposed to be able to do something like this? Compiling in release seems to be giving me good results, but it could be a silent failure.

edit: fixed wrong code

edit: this post is incorrect

Not to reply to my own message, but I wrote a minimal sample app that basically did the pseudo code above, and got the same error (using CUDA_SAFE_CALL() from the nvidia SDK)

So unless this is a mistaken error, I’m assuming that you can’t do it.

If someone from NVidia or with better hardware knowledge would like to comment it might be helpful to have clarified.

I’m sure device pointer arithmetic is valid in device functions. So i guess it is also valid within host code.

I think the mistake is, that your d_array is of type float*. If so, you should remove “sizeof(float)” from the product “(sizeof(float) * array1Size)” when adding to the device pointer.

Yeah, thank you for pointing this out! That was the problem a very simple mistake - too many years of working with void* and char* got me used to throwing the sizeof multiple in thered :)

To clarify my previous posts, you can in fact do this kind of pointer arithmetic. To avoid the mistake I made, just make sure you check the type of the pointer you’re manipulating and offset (or not) appropriately.