Single-Stepping GPU Code
TotalView allows you to single-step GPU code just like normal host code, but note that a single-step operation steps the entire warp associated with the GPU focus thread. So, when focused on a CUDA thread, a single-step operation advances all of the GPU hardware threads in the same warp as the GPU focus thread.
To advance the execution of more than one warp, you may either:
• set a breakpoint and continue the process
• select a line number in the source pane and select "Run To".
Execution of more than one warp also happens when single-stepping a __syncthreads() thread barrier call. Any source-level single-stepping operation runs all of the GPU hardware threads to the location following the thread barrier call.
Single-stepping an inlined function (nested or not) in GPU code behaves the same as single-stepping a non-inlined function. You can:
• step into an inlined function,
• step over an inlined function,
• run to a location inside an inlined function,
• single-step within an inlined function, and
• return out of an inlined function.