CUDA内核中的while循环失败

Question

我正在使用 GPU 进行一些计算来处理单词。 最初，我使用一个块（有 500 个线程）来处理一个单词。 要处理 100 个单词，我必须在 main 函数中循环内核函数 100 次。

for (int i=0; i<100; i++)
    kernel <<< 1, 500 >>> (length_of_word); 

我的内核函数如下所示：

__global__ void kernel (int *dev_length)
{
   int length = *dev_length;
   while (length > 4)
   {   //do something;
          length -=4;
   }
}

现在我想同时处理所有 100 个单词。

每个块仍将有 500 个线程，并处理一个字（每个块）。

dev_totalwordarray：存储单词的所有字符（一个接一个）

dev_length_array：存储每个单词的长度。

dev_accu_length：存储单词的累积长度（之前所有单词的总字符）

dev_salt_ 是一个大小为 500 的数组，存储无符号整数。

因此，在我的主要功能中，我有

   kernel2 <<< 100, 500 >>> (dev_totalwordarray, dev_length_array, dev_accu_length, dev_salt_);

填充 cpu 数组：

    for (int i=0; i<wordnumber; i++)
    {
        int length=0;
        while (word_list_ptr_array[i][length]!=0)
        {
            length++;
        }

        actualwordlength2[i] = length;
    }

从 cpu -> gpu 复制：

    int* dev_array_of_word_length;
    HANDLE_ERROR( cudaMalloc( (void**)&dev_array_of_word_length, 100 * sizeof(int) ) );
    HANDLE_ERROR( cudaMemcpy( dev_array_of_word_length, actualwordlength2, 100 * sizeof(int),

我的函数内核现在看起来像这样：

__global__ void kernel2 (char* dev_totalwordarray, int *dev_length_array, int* dev_accu_length, unsigned int* dev_salt_)
{

  tid = threadIdx.x + blockIdx.x * blockDim.x;
  unsigned int hash[N];

  int length = dev_length_array[blockIdx.x];

   while (tid < 50000)
   {
        const char* itr = &(dev_totalwordarray[dev_accu_length[blockIdx.x]]);
        hash[tid] = dev_salt_[threadIdx.x];
        unsigned int loop = 0;

        while (length > 4)
        {   const unsigned int& i1 = *(reinterpret_cast<const unsigned int*>(itr)); itr += sizeof(unsigned int);
            const unsigned int& i2 = *(reinterpret_cast<const unsigned int*>(itr)); itr += sizeof(unsigned int);
            hash[tid] ^= (hash[tid] <<  7) ^  i1 * (hash[tid] >> 3) ^ (~((hash[tid] << 11) + (i2 ^ (hash[tid] >> 5))));
            length -=4;
        }
        tid += blockDim.x * gridDim.x;
   }
}

但是，kernel2 似乎根本不起作用。

似乎是 while (length > 4) 造成的。

有人知道为什么吗？谢谢。

Answer 1

我不确定while 是否是罪魁祸首，但我在您的代码中看到一些让我担心的地方：

• 您的内核不产生任何输出。优化器很可能会检测到这一点并将其转换为空内核
• 几乎在任何情况下，您都不希望为每个线程分配数组。这会消耗大量内存。您的hash[N] 表将按线程分配并在内核结束时丢弃。如果N 很大（然后乘以线程总数），您可能会耗尽 GPU 内存。更不用说，访问hash 几乎和访问全局内存一样慢。
• 块中的所有线程都将具有相同的itr 值。是有意的吗？
• 每个线程仅在其自己的hash 表副本中初始化一个字段。
• 我看到hash[tid] 其中tid 是一个全局索引。请注意，即使将 hash 设为全局，您也可能会遇到并发问题。并非网格中的所有块都会同时运行。虽然一个块将初始化 hash 的一部分，但另一个块甚至可能不会启动！