I just spent two hours investigating the assert in the following code:
#include <assert.h>
__device__ int devarray[128];
extern "C" void runTest()
{
int zerobuf[128];
memset(zerobuf, 0, sizeof(zerobuf));
cudaError_t r = cudaMemcpyToSymbol(devarray, zerobuf, sizeof(zerobuf));
assert(cudaSuccess == r);
int out[128];
r = cudaMemcpyFromSymbol((void *) out, devarray, sizeof(out), cudaMemcpyDeviceToHost);
assert(cudaSuccess == r);
assert(memcmp(out, zerobuf, sizeof(out)) == 0);
}
When I compile and run this code (with a .cpp main() that calls runTest(), of course), it asserts. The zero-filled memory that was copied to the device is inexplicably not the same as the memory that is copied back from the device.
If I remove the last parameter to cudaMemcpyFromSymbol (‘kind’), it does not assert.
I thought so too… I just checked the CUDA reference manual, and was surprised to see that it’s used for device variables too. The programming guide only gives a constant example though.
That is indeed very strange. I cannot remember seeing this in any of the examples before. Also, I would expect cudaMemcpyFromSymbol does not need cudaMemcpyDeviceToHost as last parameter.
According to the Reference (I’ve never used these myself) you’re missing the offset parameter for the second call. The full prototype isn’t given, but I’m guessing the offset and the memcpyKind are optional parameters and your typo didn’t trip any syntax errors.