2

我收到一个 CUDA_EXCEPTION_5, Warp Out-of-range Address 错误，我试图找出可能导致这种情况的各种情况。

我正在将一个 C 项目（由其他人编写）移植到 CUDA。C 代码非常多寄存器，在堆栈中实例化了许多数组。我假设寄存器溢出很可能会发生，这可能会触发扭曲超出范围错误。

请注意，我想让它首先运行，然后我将开始优化代码。

我正在使用 Compute Capable 3.0 硬件，根据维基百科，它有 512KB 的“每个线程的本地内存”。我在别处读到它每个 SM 有 512KB 的寄存器空间。每个正在运行的线程是否可以有 512KB 的寄存器空间？

我目前正在执行我的内核如下（是的，我知道它超慢）：

dim3 grid(28800,1);
cuPlotLRMap<<<grid,1>>>(...)

一些细节（我不知道这会有多大帮助）：

我的硬件有 7 个 SM。有 112 个运行块，这是否意味着每个块获得 512k 的寄存器空间的 1/16？

我也明白如果一个线程超出了寄存器空间，它可能会溢出到全局内存中。发生这种情况时，并发线程是否有可能溢出到同一个全局内存空间中？

4

1 回答 1

2

“每个线程的本地内存”为 512KB。我在别处读到它每个 SM 有 512KB 的寄存器空间。每个正在运行的线程是否可以有 512KB 的寄存器空间？

请参阅CUDA C 编程指南中的计算能力表。计算能力 2.x 及更高版本的设备支持每个线程最多 512KB 的本地内存。函数 cudaDeviceSetLimit( cudaLimitStackSize, bytesPerThread ) 可用于设置值。我相信默认值是每个线程 2 KB。

我的硬件有 7 个 SM。有 112 个运行块，这是否意味着每个块获得 512k 的寄存器空间的 1/16？

计算能力 3.x 设备每个多处理器最多可以有 16 个驻留块。这假设您的寄存器/线程、线程/块或共享内存/块不会将内核限制为小于设备最大值。Visual Profiler 和 Nsight VSE CUDA Profiler 您的内核使用的配置。

目前，您只启动 1 个线程/块。您应该在每个块 (32) 中启动多个 WARP_SIZE。

我也明白如果一个线程超出了寄存器空间，它可能会溢出到全局内存中。发生这种情况时，并发线程是否有可能溢出到同一个全局内存空间中？

在编译或 JIT 时，编译器将执行寄存器分配。如果每个线程没有足够的寄存器，那么编译器将溢出到本地内存。此操作是确定性的，而不是在运行时确定的。

计算能力 3.0 设备限制为 63 个寄存器/线程。计算能力 3.5 设备限制为每个线程 255 个寄存器。

于 2012-12-03T20:57:48.483 回答