0

This is a follow up Q from cudaMemcpyToSymbol vs cudaMemcpy

I'm under the impression the -G flag is for making cudaGDB builds. I'm seeing it clear errors as well. (The error in the code below is intentional. The Q is about why the return error didn't show it.)

__constant__ float flt[480];   // 1920 bytes
__constant__ int  ints[160];   // 640 bytes

Main() {
 float* pFlts;
 cudaMalloc((void**)&pFlts, 1920+640);

 cudaError_t eerr=cudaMemcpyToSymbol(ints,pFlts,sizeof(ints),sizeof(ints),cudaMemcpyDeviceToDevice);
 printf("ErrVal= %d\n",(int)eerr);
}

When I build w/: nvcc junk.cu -o junk, and run it w/: ./junk, my result is: ErrVal= 11, which is "invalid argument".

When I build w/: nvcc junk.cu -G -o junk, and run it w/: ./junk, my result is: ErrVal= 0, which is cudaSuccess.

I don't have a lot of experience w/ the -G flag, but this seems like odd behavior. Why does -G erase return errors?

Community
  • 1
  • 1
Doug
  • 2,783
  • 6
  • 33
  • 37
  • When I comment out the first line (remove flt[480]), the error stays (11) w/ and w/o the -G. I'm thinking there is a length/size calculation issue of the symbol. And it changes w/ the -G flag. Can anyone else reproduce this? – Doug Mar 13 '13 at 19:26
  • I can reproduce it. which cuda version are you using? CUDA 5? – Robert Crovella Mar 13 '13 at 19:37
  • yes, CUDA 5.0 V0.2.1221 – Doug Mar 13 '13 at 19:39
  • When I swap the order of the first 2 lines, ErrVal= 0, w/ or w/o the -G flag. This is pretty strange. – Doug Mar 13 '13 at 19:42
  • You are again copying past the end of `ints`. Check your offset argument. – tera Mar 13 '13 at 19:47
  • And there is no guarantee that wrong addresses or offsets will be caught. – tera Mar 13 '13 at 19:48
  • @tera Please explain the 'no guarantee'. It is an obvious error that should be easy to detect. You are suggesting the range-checking is inconsistent? – Doug Mar 13 '13 at 19:53
  • 2
    I don't see this as obvious. Catching this kind of mistakes would at least require keeping range information with every symbol, something that C traditionally doesn't do. If at all, errors are mostly caught when accesses cross page borders. – tera Mar 13 '13 at 20:04
  • My prior experience w/ Memcpy lead me to believe it performed range-checking. (I thought it was odd to have range-checking in this operation.) Now I'm thinking, I probably wasn't seeing a range-fault, but rather a mem access fault. (Which makes more sense for this operation.) – Doug Mar 13 '13 at 20:57

0 Answers0