cuda - 在 CUDA 中启动的最有效块数？

Question

• 我有一个非常大的N0元素数组。
• 每个线程将循环并操作m元素。
• 我TBP每个块都有固定的线程。
• CUDA 限制每个网格的块BPG < 65535 =: BPG_max

N0 = 90现在，让我们缩小并考虑一个元素数组TBP = 32。

• 我可以开火，3 blocks of 32 threads each looping once (m = 1)这意味着3 x 32 x 1 = 96元素可能已经被操作——即浪费了 6。
• 或者我可以开火，2 blocks of 32 with m = 2这意味着2 x 32 x 2 = 128可以对元素进行操作，这是对 38 的浪费。

对于大型数组 (100MB+) 和大量循环 (10,000+)，因素会变得更大，因此浪费会变得非常大，那么如何最大限度地减少浪费？也就是说，我想要一个优化的程序（其中N表示实际完成的工作）：

Answer 1

我不会担心“浪费”线程——GPU 线程是轻量级的。

您实际上可能想要增加块大小，因为这可能会增加 GPU 的占用率。请注意，SMX（在 GeForce 6xx 系列中）只能执行 16 个并发块。使块更大将允许您调度更多线程以隐藏内存访问延迟。

Answer 2

根据您在内核中所做的一切，答案可能会或可能不会像您引用的优化问题那么简单。例如，如果您将遇到延迟问题、线程相互等待完成等问题，那么还有更多问题需要考虑。

这个网站有一些很棒的启发式方法。一些一般亮点：

为每个网格选择块

• 每个网格的块数应 >= 多处理器数。
• 在你的内核中使用的__syncthreads()越多，块就越多（这样一个块可以运行而另一个等待同步）

选择每个块的线程

• 经纱尺寸的倍数（即通常为 32）

• 通常最好选择线程数，这样每个块的最大线程数（基于硬件）是线程数的倍数。例如，最大线程数为 768，每个块使用 256 个线程往往比 512 更好，因为多个线程可以在一个块上同时运行。

• 想想你的线程是否会共享内存，如果是，你想要共享多少。

Answer 3

这实际上是一个相当复杂的问题，我怀疑是否存在 O(1) 解决方案。但我猜你可以在 CPU 上负担一些线性时间来计算最小值。

这是 wolfram alpha 的意见。