0

以前の投稿の 1 つで、カーネル機能を改善する方法を尋ねました。カーネルは、2 つの等しいサイズの行列の対応する行の間の 2 乗ユークリッド距離を計算します。エリックは、行ごとに 1 つのスレッド ブロックを使用し、その後、並列削減を適用するという非常に優れたヒントを提供しました。詳細に進む前に、以前の投稿をこれ以上複雑にしたくなかったので、この投稿を作成しました。Eric に感謝します。以下に、正しい結果が得られない .cu コードを添付しました。

__global__ void cudaEuclid( float* A, float* B, float* C, int rows, int cols )
{
    extern __shared__ float sdata[];

    unsigned int tid = threadIdx.x;
    unsigned int c = blockDim.x * blockIdx.x + threadIdx.x; // rows
    unsigned int r = blockDim.y * blockIdx.y + threadIdx.y; // cols


    sdata[ tid ] = ( A[ r*cols + c ] - B[ r*cols + c ] ) * ( A[ r*cols + c ] - B[ r*cols + c ] );

    __syncthreads();

    for ( unsigned int s = 1; s < blockDim.x; s*=2 ){
        if ( tid % (2*s) == 0 ){
            sdata[ tid ] += sdata[ tid + s ];
        }
    }
    __syncthreads();

    if ( tid == 0) C[blockIdx.x]=sdata[0];  
}

コードはhttp://developer.download.nvidia.com/compute/cuda/1.1-Beta/x86_website/projects/reduction/doc/reduction.pdfに基づいています。最適化されたバージョンではありません。基本的なポイントをキャッチしたいだけです。初期化するところに問題があると思いますsdata。また、カーネルの初期化は次の方法で行われます。

int threadsPerBlock = 256;  
int blocksPerGrid = ceil( (double) numElements  / threadsPerBlock);

dim3 dimBlock(1, threadsPerBlock); 
dim3 dimGrid(blocksPerGrid, 1); 

cudaEuclid<<<dimGrid, dimBlock>>>( d_A, d_B, d_C, rows, cols );

ありがとうございます。私の無知をお詫び申し上げます。

4

2 に答える 2

0

sdata[tid] += sdata[tid]; ==>同じ値を2回追加する必要があります

sdata[tid] += sdata[tid +s]

于 2013-10-03T19:10:45.643 に答える