CUDA-GDB(5)——内核聚焦

安全二次方(security²)博客目录导读

目录

5.1. 软件坐标与硬件坐标

5.2. 当前焦点

5.3. 切换焦点

---

一个CUDA应用程序可能运行多个主机线程和许多设备线程。为了简化应用程序状态信息的可视化，命令将应用于当前焦点实体。

当焦点设置到主机线程时，命令将仅适用于该主机线程（除非应用程序完全恢复运行）。在设备端，焦点始终设置为最细粒度级别——设备线程。

5.1. 软件坐标与硬件坐标

设备线程属于一个块，而块又属于一个内核。线程、块和内核是焦点的软件坐标。设备线程在通道（lane）上运行。通道属于一个线程束，线程束属于一个流式多处理器(SM)，而SM又属于一个设备。lane、线程束、SM和设备是焦点的硬件坐标。只要保持一致性，软件坐标和硬件坐标可以互换使用或同时使用。

有时会使用另一种软件坐标：网格(grid)。网格与内核(kernel)的区别在于作用范围。网格ID在单个GPU内是唯一的，而内核ID在所有GPU之间都是唯一的。因此，内核与(网格,设备)元组之间存在一一对应的映射关系。

注意

如果启用了软件抢占功能（set cuda software_preemption on），设备线程恢复执行时，对应的硬件坐标可能会发生变化。但软件坐标将保持不变，在设备线程的整个生命周期内都不会改变。

5.2. 当前焦点

要检查当前焦点，请使用cuda命令后跟感兴趣的坐标：

(cuda-gdb) cuda device sm warp lane block thread
block (0,0,0), thread (0,0,0), device 0, sm 0, warp 0, lane 0
(cuda-gdb) cuda kernel block thread
kernel 1, block (0,0,0), thread (0,0,0)
(cuda-gdb) cuda kernel
kernel 1

5.3. 切换焦点

要切换当前焦点，使用cuda命令后跟需要更改的坐标：

(cuda-gdb) cuda device 0 sm 1 warp 2 lane 3
[Switching focus to CUDA kernel 1, grid 2, block (8,0,0), thread
(67,0,0), device 0, sm 1, warp 2, lane 3]
374 int totalThreads = gridDim.x * blockDim.x;

如果命令未完全定义指定的焦点，调试器将假定省略的坐标设置为当前焦点中的坐标，包括块和线程的子坐标。

(cuda-gdb) cuda thread (15)
[Switching focus to CUDA kernel 1, grid 2, block (8,0,0), thread
(15,0,0), device 0, sm 1, warp 0, lane 15]
374 int totalThreads = gridDim.x * blockDim.x;

块和线程参数的括号是可选的。

(cuda-gdb) cuda block 1 thread 3
[Switching focus to CUDA kernel 1, grid 2, block (1,0,0), thread (3,0,0),
device 0, sm 3, warp 0, lane 3]
374 int totalThreads = gridDim.x * blockDim.