How can I debug this CUDA / OptiX crash with cuda-gdb?

I have a repeatable crash in my code, so I tried to catch it in cuda-gdb but it doesn’t seem to be able to get into the device code, even though everything is build with ‘-g’ include *.cu with nvcc. My issue has to do with math at a certain point, so I would really like to get in the device code to look at variables at time of crash if that is possible. Below is all the information I can get currently.

--Type <RET> for more, q to quit, c to continue without paging--

CUDA Exception: Warp Illegal Address
The exception was triggered at PC 0x555557bc2e60

Thread 1 "SensorApp" received signal CUDA_EXCEPTION_14, Warp Illegal Address.
[Switching focus to CUDA kernel 33, grid 296, block (395,6,0), thread (62,0,0), device 0, sm 10, warp 13, lane 30]
0x0000555557bc2f00 in __closesthit__radiance_ptID_0xe1371196ca236d47<<<(832,13,1),(64,1,1)>>> ()

You need to build with -g -G to debug CUDA device code. For questions about how to debug optix, I suggest posting those on the optix forum.

You likely won’t be able to inspect variables in device memory after the crash. I would suggest understanding the point of the crash, then setting a breakpoint before that point, perhaps using conditional breakpoints.

A general overview of CUDA debugging is in session 12 here.