嵌套for循环中的Cuda减少

Cuda reduction in nested for loops

本文关键字：Cuda 减少 for 循环 嵌套      更新时间：2023-10-16

我有一个关于CUDA中某种约简的问题。

distance是具有gridSize*numberOfAngles元素的矩阵，fftData是具有numberOfAngles*NFFT元素的矩阵。grid_magnitude是我想要存储计算结果的结果矩阵，它具有gridSize元素。

我想计算fftData中对应于distance中一个特定值的索引。之后，fftData中该索引处的值被添加到对应的gridPoint处的grid_magnitude。

这是我迄今为止的代码：

__global__ void calcBackProjection(cuFloatComplex* fftData, 
                                   cuFloatComplex* grid_magnitude,
                                   float* distance,
                                   int gridSize,
                                   int numberOfAngles,
                                   float rangeBin,
                                   int NFFT)
{
 int gridPointIdx = threadIdx.x + blockIdx.x*blockDim.x;
 while(gridPointIdx < gridSize)
 {
    for(int angleIdx = 0; angleIdx < numberOfAngles; angleIdx++)
    {       
        //find rangeBin in fftData corresponding to distance
        float curDistance = distance[gridPointIdx + angleIdx*gridSize];
        int wantedIdx = floor(curDistance / rangeBin);
        if(wantedIdx < NFFT)
            {                                   
                grid_magnitude[gridPointIdx + angleIdx*gridSize] =  
              addCmplx(fftData[wantedIdx + angleIdx*NFFT], grid_magnitude[gridPointIdx +     
                angleIdx*gridSize]);
            }                   
    }
    gridPointIdx += blockDim.x * gridDim.x;     
 }   
}

gridPointIdx对于每个线程应该是唯一的，因此每个线程应该在grid_magnitude中的不同位置进行写入。但这似乎不起作用，因为grid_magnitude没有任何更改。

我错过了什么？

我没能在完全并行的二维索引中做到这一点，也许我应该使用共享内存，但我如何将网格大小划分为线程部分使用？

我把代码改了一小部分。

__global__ void calcBackProjection(cuFloatComplex* fftData, cuFloatComplex* grid_magnitude,
float* distance, int gridSize, int numberOfAngles, float rangeBin, int NFFT){
int gridPointIdx = threadIdx.x + blockIdx.x*blockDim.x;
while(gridPointIdx < gridSize){
    for(int angleIdx = 0; angleIdx < numberOfAngles; angleIdx++){       
        float curDistance = distance[gridPointIdx + angleIdx*gridSize];
        int wantedIdx = ceil(curDistance / rangeBin) - 1;
        if(wantedIdx){
            int fftIdx = wantedIdx + angleIdx*NFFT;
            int gridIdx=  gridPointIdx + angleIdx*gridSize;
            if((fftIdx < NFFT*numberOfAngles) && (gridIdx < gridSize*numberOfAngles)){                  
                grid_magnitude[gridIdx] = cuCaddf(fftData[fftIdx], grid_magnitude[gridIdx]);
            }
        }
    }
    gridPointIdx += blockDim.x * gridDim.x;     }}

问题似乎是，编译器没有计算变量curDistance和wantedIdx。当我想知道这些值时，它会说"在目标位置没有值"。这似乎是在grid_magnitude[gridIdx] = cuCaddf(fftData[fftIdx], grid_magnitude[gridIdx]);检测到访问违规的原因我看了一些关于这个问题的其他答案，比如这里和这里，但这些对我帮助不大。

相关文章：