How to use cuda-gdb python debugging?

Question

I write a hello.py that is simply

print("hi")

and then run

cuda-gdb python3 hello.py

I get:

Reading symbols from python3...
(No debugging symbols found in python3)
"/home/x/Desktop/py_projects/hello.py" is not a core dump: file format not recognized

How to debug if I call cuda functions in python code?

Answer 1

Its possible to use cuda-gdb from python, assuming you only need to debug the C/C++ portion. I don't know of a debugger that can jump from debugging python to debugging CUDA C++. Here is one possible approach, a copy of what is presented here.

To debug a CUDA C/C++ library function called from python, the following is one possibility, inspired from this article.

1. For this walk through, I will use the t383.py and t383.cu files verbatim from this answer, and I'll be using CUDA 10, python 2.7.5, on CentOS7

2. Compile your CUDA C/C++ library using the -G and -g switches, as you would to do ordinary debug:

$ nvcc -Xcompiler -fPIC -std=c++11 -shared -arch=sm_60 -G -g -o t383.so t383.cu -DFIX

3. We'll need two terminal sessions for this. I will refer to them as session 1 and session 2. In session 1, start your python interpreter:

$ python
...
>>>

4. In session 2, find the process ID associated with your python interpreter (replace USER with your actual username):

$ ps -ef |grep USER
...
USER    23221 22694  0 23:55 pts/0    00:00:00 python
...
$

In the above example, 23221 is the process ID for the python interpreter (use man ps for help)

5. In session 2, start cuda-gdb so as to attach to that process ID:

$ cuda-gdb -p 23221
... (lots of spew here)
(cuda-gdb)

6. In session 2, at the (cuda-gdb) prompt, set a breakpoint at a desired location in your CUDA C/C++ library. For this example, we will set a breakpoint at one of the first lines of kernel code, line 70 in the t383.cu file. If you haven't yet loaded the library (we haven't, in this walk through), then cuda-gdb will point this out and ask you if you want to make the breakpoint pending on a future library load. Answer y to this (alternatively, before starting this cuda-gdb session, you could have run your python script once from within the interpreter, as we will do in step 7 below. This would load the symbol table for the library and avoid this prompt). After the breakpoint is set, we will issue the continue command in cuda-gdb in order to get the python interpreter running again:

(cuda-gdb) break t383.cu:70
No symbol table is loaded.  Use the "file" command.
Make breakpoint pending on future shared library load? (y or [n]) y
Breakpoint 1 (t383.cu:70) pending.
(cuda-gdb) continue
Continuing.

7. In session 1, run your python script:

>>> execfile("t383.py")
init terrain_height_map
1, 1, 1, 1, 1,
1, 1, 1, 1,
1, 1, 1, 1,

8. our python interpreter has now halted (and is unresponsive), because in session 2 we see that the breakpoint has been hit:

[New Thread 0x7fdb0ffff700 (LWP 23589)]
[New Thread 0x7fdb0f7fe700 (LWP 23590)]
[Switching focus to CUDA kernel 0, grid 1, block (0,0,0), thread (0,0,0), device 0, sm 0, warp 0, lane 0]

Thread 1 "python" hit Breakpoint 1, update_water_flow<<<(1,1,1),(1024,1,1)>>> (
    water_height_map=0x80080000800000, water_flow_map=0xfffcc000800600,
    d_updated_water_flow_map=0x7fdb00800800, SIZE_X=4, SIZE_Y=4) at t383.cu:70
70          int col = index % SIZE_X;
(cuda-gdb)

and we see that the breakpoint is at line 70 of our library (kernel) code, just as expected. ordinary C/C++ cuda-gdb debug can proceed at this point within session 2, as long as you stay within the library function.

When you are finished debugging (you may need to remove any breakpoints set) you can once again type continue in session 2, to allow control to return to the python interpreter in session 1, and for your application to finish.

Answer 2

To complete Robert's answer, if you are using CUDA-Python, you can use option --args in order to pass a command-line that contains arguments. For example, this is a valid command-line:

$ cuda-gdb --args python3 hello.py

Your original command is not valid because, without --args, cuda-gdb takes in parameter a host coredump file.

Here is the complete command line with an example from the CUDA-Python repository:

$ cuda-gdb -q --args python3 simpleCubemapTexture_test.py
Reading symbols from python3...
(No debugging symbols found in python3)
(cuda-gdb) set cuda break_on_launch application
(cuda-gdb) run
Starting program: /usr/bin/python3 simpleCubemapTexture_test.py.
...
[Switching focus to CUDA kernel 0, grid 1, block (0,0,0), thread (0,0,0), device 0, sm 0, warp 0, lane 0]
0x00007fff67858600 in transformKernel<<<(8,8,1),(8,8,1)>>> ()
(cuda-gdb) p $pc
$1 = (void (*)()) 0x7fff67858600 <transformKernel>
(cuda-gdb) bt
#0  0x00007fff67858600 in transformKernel<<<(8,8,1),(8,8,1)>>> ()