CUDA:sum[0] 的值随每次执行而变化

Question

 __global__ void gpu_Heat (float *h, float *g, float * sum, int N) {
        int nbx, bx, nby, by;
        float diff = 0.0;
        nbx = (N-2)/blockDim.x;
        bx = nbx/gridDim.x;
        nby = (N-2)/blockDim.y;
        by = nby/gridDim.y;     
        unsigned int ii = blockIdx.x*blockDim.x+threadIdx.x;
        unsigned int jj = blockIdx.y*blockDim.y+threadIdx.y;
        unsigned int jid = (ii)*(N-2)+(jj);
         for (int i=1+ii*bx; i<=min((ii+1)*bx, N-2); i++) 
                    for (int j=1+jj*by; j<=min((jj+1)*by, N-2); j++) {
                    g[i*N+j]= 0.25 * (h[ i*N + (j-1)]+ 
                                 h[ i*N +(j+1) ]+
                                 h[ (i-1)*N + j]+ 
                                 h[ (i+1)*N + j]); 
                        diff = g[i*N+j] - h[i*N+j];
                        sum[(i-1)*(N-2)+(j-1)] = diff * diff;
                }
        __syncthreads();
        for(unsigned int s=((N-2)*(N-2))/2; s>0; s>>=1){    
            if(jid<s){
                sum[jid]+=sum[jid+s];           
            }
            __syncthreads();    

        }
    }

所以我的问题是 sum[0] 的值包含并行归约的最终结果，每次执行该内核时都会发生变化，即使输入是相同的而且我不知道我在做什么错误的。如果在 CPU 中减少相同的求和矩阵，则执行很好，但在 GPU 中并行减少会给我带来问题。

    dim3 Grid = (16,16);
    dim3 Block = (16,16);
    gpu_Heat<<<Grid,Block>>>(dev_u, dev_uhelp, dev_sum, np);
    cudaThreadSynchronize();                        // wait for all threads to complete
    cudaErrorCheck(cudaMemcpy(param.u,dev_u,np*np*sizeof(float),cudaMemcpyDeviceToHost));
    cudaErrorCheck(cudaMemcpy(param.uhelp,dev_uhelp,np*np*sizeof(float),cudaMemcpyDeviceToHost));
    cudaErrorCheck(cudaMemcpy(sum,dev_sum,sum_size*sizeof(float),cudaMemcpyDeviceToHost));

我在此处明确显示了由我正在使用的测试中的代码计算的 block 和网格参数。谢谢解答

Answer 1

您正在启动具有多个 block 的 GPU 内核。虽然一个 block 中的线程仍在第一个 for 循环中计算它们的 sum，但其他一些 block 可能在第二个 for 循环中进行并行归约.这两个 for 循环具有数据依赖性。由于流式多处理器上的 block 调度发生在幕后，并且每次运行可能会有所不同，因此每次都会得到不同的结果。 __syncthreads(); 在 for 循环之间同步块(synchronized block)内的线程，但没有用于多个 block 之间同步的机制或指令，除非返回到主机并发布另一个内核。

在您的情况下，即使您简单地将 for 循环分开，您得到的结果仍然可能是错误的，因为您的减少发生在多个 block 中，并且 block 的调度再次不确定。