How to debug into CUDA kernel code using visual studio 2008?

Question

Hey, I am using Visual Studio 2008, with CUDA 3.2. I am trying to debug into a function with this signature:

 MatrixMultiplication_Kernel<<<dimGrid, dimBlock>>>(Md, Nd, Pd, Width);

I can step into the function, however when I get into the function it doesn't let me step over any of the code and tells me that no source is available. Anybody knows how to debug into this properly?

Thank you!

Answer 1

You can't debug CUDA kernel code using visual studio. Visual studio can only debug programs that run on the CPU.

In fact, I don't think that you can debug CUDA kernel code in Windows at all any more. CUDA used to have a host emulation mode, but that was removed in 3.0. The only debugging tools available are cuda-gdb and ocelot, and, as far as I know, neither of them supports Windows.

Answer 2

Are you using Nsight? Vanilla Visual Studio cannot step into device code, but with Nsight, this is possible. Unfortunately, limitations of the Windows device driver model mean there are some serious restrictions. It used to be that two machines were required to do the debugging. The target machine would run the CUDA code on it's GPU, and had to be using the TCC driver. The host machine would run Visual Studio and control the target. It seems that it is now possible to debug CUDA code on the same machine as long as you have two GPU's, one for compute and one for display. There are a number of other problems mentioned in the user guide.

This may not be an acceptable answer, but: If you can compile your code on Linux, you can use cuda-gdb to debug kernels. cuda-gdb is part of the Linux CUDA toolkit.

Answer 3

To debug device code within Visual Studio you will need Parallel Nsight. The standard version is free and offers the debugging you require as well as device code profiling.

If you want to debug on a single machine then you will need two GPUs (since the GPU running the code will be stopped when it hits a breakpoint, and hence your display would block as well). They don't need to be high-end GPUs though, anything from G92 onwards will do (including most Fermi GPUs as listed here).