On Fermi, my coda C code performs well, however it gets me wrong results on Kepler. So I cuda-memcheck the code and got the error message:
========= CUDA-MEMCHECK
========= Invalid __shared__ read of size 8
========= at 0x00000868 in datTime(int*, int*, double*, double const *, double const *, double const *, double const *, double const *, double const *, double const *, double const *, int, int, int)
========= by thread (127,0,0) in block (62,0,0)
========= Address 0x00000878 is out of bounds
========= Saved host backtrace up to driver entry point at kernel launch time
========= Host Frame:/usr/lib64/libcuda.so (cuLaunchKernel + 0x2c5) [0x14ad95]
========= Host Frame:rr [0x1c968]
========= Host Frame:rr [0x3be33]
========= Host Frame:rr [0x3c0e]
========= Host Frame:rr [0x3a09]
========= Host Frame:rr [0x3ac0]
========= Host Frame:rr [0x3473]
========= Host Frame:/lib64/libc.so.6 (__libc_start_main + 0xfd) [0x1ed1d]
========= Host Frame:rr [0x27a9]
=========
========= Program hit error 30 on CUDA API call to cudaThreadSynchronize
========= Saved host backtrace up to driver entry point at error
========= Host Frame:/usr/lib64/libcuda.so [0x2ef673]
========= Host Frame:rr [0x3a8f6]
========= Host Frame:rr [0x3478]
========= Host Frame:/lib64/libc.so.6 (__libc_start_main + 0xfd) [0x1ed1d]
========= Host Frame:rr [0x27a9]
Any suggestions?
Thanks in advance.
Was your code pre-compiled, or did you recompile on the new system? In case of the latter, it may be a matter of compiler options.
Invalid __shared__ read of size 8
Address 0x00000878 is out of bounds
It looks a "read out of bonds" from the shared memory. I suggest you to verify how you (programmatically) set the balance between shared memory and cache memory.
Hello, Qingjun! This is a very intriguing issue. It may be a problem due many possibilities. I think you should try to run your code on another graphics card from another PC. If the error does not persist, I suggest you to change your graphics card because it may be damaged.
Best wishes and good luck with your research!
Hi, Lauro de Paula, Marco Salvatore Nobile, and Allard Jan Van Marle, thanks for your suggestions! After trying to run on another graphics card and checking the compiler options when recompiling the code that runs well one C1060 tesla T10, it persists on kepler K20 and K10. The problem has been solved with the help of an NVIDIA export: explicitly specify Ns (the amount of shared memory) when launching kernel on Kepler. It is funny that the same code without specifying Ns ran well on C1060 tesla T10. Thanks again!
- Badges
- Science topic