我应该检查内核代码中的线程数吗？答案

【问题标题】：Should I check the number of threads in kernel code?我应该检查内核代码中的线程数吗？
【发布时间】：2018-01-07 18:21:29
【问题描述】：

我是 CUDA 的初学者，我的同事总是使用以下包装设计内核：

__global__ void myKernel(int nbThreads)
{
    int threadId = blockDim.x*blockIdx.y*gridDim.x  //rows preceeding current row in grid
            + blockDim.x*blockIdx.x             //blocks preceeding current block
            + threadIdx.x;

    if (threadId < nbThreads)
    {
        statement();
        statement();
        statement();
    }
}

他们认为在某些情况下 CUDA 可能会启动比指定更多的线程来进行对齐/扭曲，因此我们需要每次都检查它。但是，到目前为止，我还没有在 Internet 上看到他们实际执行此验证的示例内核。

CUDA 真的可以启动比指定块/网格尺寸更多的线程吗？

【问题讨论】：

标签： cuda

【解决方案1】：

CUDA 不会启动超过块/网格尺寸指定的线程数。

但是，由于块尺寸的粒度（例如，块尺寸最好是 32 的倍数，并且大小限制为 1024 或 512），通常情况下很难匹配线程网格在数值上等于所需的问题大小。

在这些情况下，典型的行为是启动更多线程，根据块粒度有效地向上舍入到下一个偶数大小，并使用内核中的“线程检查”代码来确保“额外线程” ，即超出问题大小的那些，不要做任何事情。

在你的例子中，这可以通过写作来澄清：

__global__ void myKernel(int problem_size)


    if (threadId < problem_size)

它传达了预期的内容，即只有与问题大小相对应的线程（可能与启动的网格大小不匹配）才能执行任何实际工作。

作为一个非常简单的例子，假设我想在一个长度为 10000 个元素的向量上进行向量相加。 10000 不是 32 的倍数，也不小于 1024，所以在典型的实现中，我会启动多个线程块来完成这项工作。

如果我希望每个线程块是 32 的倍数，那么我可以选择没有多少线程块可以给我正好 10000 个线程。因此，我可能会在一个线程块中选择 256 个线程，并启动 40 个线程块，总共给我 10240 个线程。使用线程检查，我可以防止“额外”的 240 个线程做任何事情。

【讨论】：