以前の投稿の 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 );
ありがとうございます。私の無知をお詫び申し上げます。