具有两个不同线程索引的 cuda 内核答案

【问题标题】：a cuda kernel with two different threads indexes具有两个不同线程索引的 cuda 内核
【发布时间】：2017-01-25 20:29:09
【问题描述】：

我正在 CUDA 中编写一个使用两个不同线程的内核。这是场景如何进行的。我有 3 个参数数组（alpha、beta、sigma），每个数组的大小为 64。我有一个大小为 10000 的 randomValue 数组。

我的目标是创建一个大小为 64 的数组，该数组计算一个函数 (NextRate)，其中对应参数的所有不同 randomValues。那，我可以通过将 tid2 定义为参数的索引并将 tid3 定义为 randomValues 的索引来按照代码中所示的方式进行操作。这是我的内核：

__global__ void evaluateVasicek(KernelArray<double> crrntMonthMdlData, KernelArray<double> crrntMonthMrktData,
                            KernelArray<double> alpha, KernelArray<double> beta,
                            KernelArray<double> sigma, KernelArray<double> nextRateRands,
                            const int NP, double r0, KernelArray<double> dr)
{
  int tid = blockIdx.x * blockDim.x + threadIdx.x;

  if (tid >= 640000)
     return;

  int tid2 = tid / 10000; // The index for Parameters
  int tid3 = tid % 10000; // The index for randomArray

  dr._array[tid2] += NextRate(nextRateRands._array[tid3], alpha._array[tid2], beta._array[tid2], sigma._array[tid2], r0);
  __syncthreads();

  if (tid3 == 0)
      printf("dr for tid %d is %f\n", tid2, dr._array[tid2] );

}

当我想总结所有 10000 博士的相应参数时，我写了以下行：

dr._array[tid2] += NextRate(nextRateRands._array[tid3], alpha._array[tid2], beta._array[tid2], sigma._array[tid2], r0);
__syncthreads();

但这似乎对我不起作用。我只是得到最后一个计算值，没有发生线程同步。

【问题讨论】：

标签： c++ cuda

【解决方案1】：

自从找到答案后，我认为让其他人也知道是件好事。

__global__ void evaluateVasicek(KernelArray<double> crrntMonthMdlData, KernelArray<double> crrntMonthMrktData,
                            KernelArray<double> alpha, KernelArray<double> beta,
                            KernelArray<double> sigma, KernelArray<double> nextRateRands,
                            const int NP, double r0, KernelArray<double> dr, KernelArray<double> dr64,
                            KernelArray<double> rNext, KernelArray<double> tau, KernelArray<double> error)
{
  int tid = blockIdx.x * blockDim.x + threadIdx.x;

  if (tid >= 640000)
  return;

  int tid2 = tid / 10000; // The index for Parameters
  int tid3 = tid % 10000; // The index for randomArray

  dr._array[tid] += NextRate(nextRateRands._array[tid3], alpha._array[tid2], beta._array[tid2], sigma._array[tid2], r0);
  __syncthreads();

  if (tid3 == 0) {

  for (int i = 0; i < 10000; ++i)
     dr64._array[tid2] += dr._array[tid2 * 10000 + i];


     rNext._array[tid2] = r0 + dr64._array[tid2] / 10000;

所以基本上我所做的就是让 dr 成为一个大小为 640000 的数组，每个线程计算一个函数调用。

然后，我创建了另一个名为 dr64 的数组，大小为 64，并且对于参数 (tid2) 的每个索引，我将所有对应的 10000 个值相加。

【讨论】：