0

我正在编写一个 CUDA 和约简代码，它采用从元素 begin_index 到 end_index 开始的数组的绝对值的总和（我正在使用一个具有可变线程数的块）。但是，当我将数组传递A给 时reduce_fabs()，所有来自A[4]和更高索引的内容最终都会以某种方式设置为零。这是代码和函数调用以及我尝试过的更多解释。

这是调用 sum-reduction 内核的内核：

__device__ void tridiag(float *A,int *dim, float *diag,float *offdiag) {

    A[0] = 1.0f; A[1] = 2.0f; A[2] = 3.0f;
    A[3] = 4.0f; A[4] = 5.0f; A[5] = 6.0f;
    diag[0] = reduce_fabs(A,0,3);
    __syncthreads();
    diag[1] = reduce_fabs(A,0,4);

    return;
}

这是减和核：

__device__ float reduce_fabs(float *v, int begin_index, int end_index) {
    extern __shared__ float sum_array[];
    int tid = threadIdx.x;
    if(tid >= begin_index && tid <= end_index) {
        sum_array[tid-begin_index] = fabs(v[tid]);
        sum_array[tid+end_index-begin_index+1] = 0;
    }
    __syncthreads();
    for(int j=1;j<=(end_index-begin_index);j*=2) {
        if((tie-begin_index)%(2*j) == 0 && tie >= begin_index && tid <= end_index) {
            sum_array[(tie-begin_index)] += sum_array[(tie-begin_index)+j];
        }
        __syncthreads();
    }

    return sum_array[0];
}

看过代码，对问题的更具体的描述是，当我将数组传递A给时reduce_fabs()，元素 4 的值变成A[4] = 0了错误的。A[0], A[1],A[2]和A[3]对任何 都很好end_index，但每end_index大于 3 都会导致超过 3 的元素A在通过归约内核时被设置为零。

这是我已经尝试过的：

  • 我试图diag[0] = A[4]确保最初的任务是有效的。它是。
  • 我消除了 的总和部分reduce_fabs()，在第一个之后停止，__syncthreads()问题仍然存在。
  • 我消除了对总和感兴趣的元素之外的元素的归零；即我评论//sum_array[tid+end_index-begin_index+1] = 0了（这部分的总结也被评论了）。没有成功。
  • 我消除if(tie >= begin_index && tid <= end_index)了，因此sum_array分配给所有线程（总和也注释掉了）无济于事。

这是运行它的主要功能：

#include<stdio.h>
#include<stdlib.h>
#include<cuda.h>

int main() {
    int n = 10;
    float *A = (float *)malloc(n*sizeof(*A));
    float *diag = (float *)malloc(n*sizeof(*A));
    float *offdiag = (float *)malloc(n*sizeof(*A));

    int *p_n;
    float *p_A, *p_diag, *p_offdiag;

    cudaMalloc((void**) &p_A,n*sizeof(float));
    cudaMalloc((void**) &p_diag,n*sizeof(float));
    cudaMalloc((void**) &p_offdiag,n*sizeof(float));
    cudaMalloc((void**) &p_n,sizeof(int));

    cudaMemcpy(p_n,&n,n*sizeof(int),cudaMemcpyHostToDevice);

    tridiag<<<1,n>>>(p_A,p_n,p_diag,p_offdiag);

    cudaMemcpy(A,p_A,n*sizeof(float),cudaMemcpyDeviceToHost);
    cudaMemcpy(diag,p_diag,n*sizeof(float),cudaMemcpyDeviceToHost);

    printf("A[0] = %f  A[1] = %f  A[2] = %f\n",A[0],A[1],A[2]);
    printf("A[3] = %f  A[4] = %f  A[5] = %f\n",A[3],A[4],A[4]);
    printf("diag[0] = %f  diag[1] = %f\m",diag[0],diag[1]);

    cudaFree(p_A);
    cudaFree(p_diag);
    cudaFree(p_offdiag);
    free(A);
    free(diag);
    free(offdiag);

    return 0;
}
4

0 回答 0