2

コードを並行して実行する（CUDA）ように一生懸命努力しています。単純化されたコードは次のとおりです。

float sum = ...         //sum = some number
for (i = 0; i < N; i++){
    f = ...             // f = a function that returns a float and puts it into f
    sum += f;
}

私が抱えている問題は、スレッド間で共有sum+=fする必要があるためです。sum ( ) を宣言するときsumに引数を使用しようとしましたが、うまくいきませんでした (適切な結果が得られませんでした)。リダクションについても聞いたことがありますが (OpenMP での使用方法も知っています)、ここでの適用方法がわかりません。__shared____shared__ float sum

どんな助けでも大歓迎です。ありがとう！

4

2 に答える 2

5

コードは次のとおりです。

#include <stdio.h>
__global__ void para(float* f, int len) {
    int i = threadIdx.x + blockDim.x * blockIdx.x;
    if (i < len ){
        // calculate f[i], f in iteration ith
        f[i] = i;
    }
}

int main(int argc, char ** argv) {
    int inputLength=1024;
    float * h_f;
    float * d_f;
    int size = inputLength*sizeof(float);

    h_f = (float *) malloc(size);
    cudaMalloc((void**)&d_f , size);
    cudaMemcpy(d_f, h_f, size, cudaMemcpyHostToDevice);

    dim3 DimGrid((inputLength)/256 +1 , 1 , 1);
    dim3 DimBlock(256 , 1, 1);

    para<<<DimGrid , DimBlock>>>(d_f , inputLength);
    cudaThreadSynchronize();

    cudaMemcpy(h_f, d_f, size , cudaMemcpyDeviceToHost);

    cudaFree(d_f);

    // do parallel reduction
    int i;
    float sum=0;
    for(i=0; i<inputLength; i++)
        sum+=h_f[i];

    printf("%6.4f\n",sum);

    free(h_f);

    return 0;
}

並列リダクション セクションは、作業中の CUDA 並列合計リダクション (たとえば、この) に置き換えることができます。すぐに私はそれを変更するのに時間がかかります。

編集：

CUDA で並列リダクションを実行するコードは次のとおりです。

#include <stdio.h>
__global__ void para(float* f, int len) {
    int i = threadIdx.x + blockDim.x * blockIdx.x;
    if (i < len ){
        // calculate f[i], f in iteration ith
        f[i] = i;
    }
}
__global__ void strideSum(float *f, int len, int strid){
    int i = threadIdx.x + blockDim.x * blockIdx.x;
    if(i+strid<len){
        f[i]=f[i]+f[i+strid];
    }
}

#define BLOCKSIZE 256
int main(int argc, char ** argv) {
    int inputLength=4096;
    float * h_f;
    float * d_f;
    int size = inputLength*sizeof(float);

    h_f = (float *) malloc(size);
    cudaMalloc((void**)&d_f , size);
    cudaMemcpy(d_f, h_f, size, cudaMemcpyHostToDevice);

    dim3 DimGrid((inputLength)/BLOCKSIZE +1 , 1 , 1);
    dim3 DimBlock(BLOCKSIZE , 1, 1);

    para<<<DimGrid , DimBlock>>>(d_f , inputLength);
    cudaThreadSynchronize();

    int i;
    float sum=0, d_sum=0;
    // serial sum on host. YOU CAN SAFELY COMMENT FOLLOWING COPY AND LOOP. intended for sum validity check.
    cudaMemcpy(h_f, d_f, size , cudaMemcpyDeviceToHost);
    for(i=0; i<inputLength; i++)
        sum+=h_f[i];

    // parallel reduction on gpu
    for(i=inputLength; i>1; i=i/2){
        strideSum<<<((i/BLOCKSIZE)+1),BLOCKSIZE>>>(d_f,i,i/2);
        cudaThreadSynchronize();
    }
    cudaMemcpy(&d_sum, d_f, 1*sizeof(float) , cudaMemcpyDeviceToHost);

    printf("Host -> %6.4f, Device -> %6.4f\n",sum,d_sum);

    cudaFree(d_f);
    free(h_f);

    return 0;
}
于 2012-12-10T06:13:10.183 に答える
1

必要なのは、数値の範囲をスレッドにマップし、各スレッドにその範囲を追加させてから、削減フェーズを行うことです。削減は、各スレッドからの合計を追加します。

于 2012-12-10T02:19:18.593 に答える