迭代第二个周期，CUDA 中的总和减少

Iteration second cycle with sum reduction in CUDA

本文关键字：第二个 周期 CUDA 迭代      更新时间：2023-10-16

>我必须将此代码从c ++并行化到CUDA C 

  for(ihist = 0; ihist < numhist; ihist++){ 
      for(iwin = 0; iwin<numwin; iwin++){
          denwham[ihist] += (numbinwin[iwin]/g[iwin])*exp(F[iwin]-U[ihist]); 
          }
          Punnorm[ihist] = numwham[ihist]/denwham[ihist];
        }

在 CUDA C 中，使用总和约简：

extern __shared__ float sdata[];
  int tx = threadIdx.x;
  int i=blockIdx.x;
  int j=blockIdx.y;
  float sum=0.0;
  float temp=0.0;
  temp=U[j];

   if(tx<numwin)
   {
    sum=(numbinwin[tx]/g[tx])*exp(F[tx]- temp); 
    sdata[tx] = sum;
     __syncthreads();  
   }

  for(int offset = blockDim.x / 2;offset > 0;offset >>= 1)
  {
   if(tx < offset)
   {
    // add a partial sum upstream to our own
    sdata[tx] += sdata[tx + offset];
   }
   __syncthreads();
  }
   // finally, thread 0 writes the result
  if(threadIdx.x == 0)
  {
   // note that the result is per-block
   // not per-thread
   denwham[i] = sdata[0];
    for(int k=0;k<numhist;k++)
    Punnorm[k] = numwham[k]/denwham[k];
  }

并通过以下方式初始化它：

 int smem_sz = (256)*sizeof(float);
  dim3 Block(numhist,numhist,1);
  NewProbabilitiesKernel<<<Block,256,smem_sz>>>(...);

我的问题是我无法使用exp迭代U，我已经尝试了以下方法：

a) loop for/while inside the kernel that iterates over U 
b) iterate by thread
c) iterate to block

所有这些尝试导致我在C++代码和代码 cuda 之间得出不同的结果。如果我放一个常量而不是 U [i]，代码工作正常！

你有什么想法可以帮助我吗？

谢谢。

看起来您可以将U移出内部循环

for(iwin = 0; iwin<numwin; iwin++){
    denwham += numbinwin[iwin] / g[iwin] * exp(F[iwin]); 
}
for(ihist = 0; ihist < numhist; ihist++){ 
    Punnorm[ihist] = numwham[ihist] / denwham * exp(U[ihist]);
}

更新

之后，您可以使用 2 个简单内核而不是 1 个复杂内核来完成任务。

  1. 将内核还原为计算denwham;
  2. 一维转换内核以计算Punnorm;
相关文章：