0
CUDA에서 내적 제품을 구현하고 결과를 MATLAB이 반환하는 것과 비교하려고합니다.원자 연산을 사용하는 CUDA의 점 제품 - 잘못된 결과 얻기
#include <stdio.h>
#define N (2048 * 8)
#define THREADS_PER_BLOCK 512
#define num_t float
// The kernel - DOT PRODUCT
__global__ void dot(num_t *a, num_t *b, num_t *c)
{
__shared__ num_t temp[THREADS_PER_BLOCK];
int index = threadIdx.x + blockIdx.x * blockDim.x;
temp[threadIdx.x] = a[index] * b[index];
__syncthreads(); //Synchronize!
*c = 0.00;
// Does it need to be tid==0 that
// undertakes this task?
if (0 == threadIdx.x) {
num_t sum = 0.00;
int i;
for (i=0; i<THREADS_PER_BLOCK; i++)
sum += temp[i];
atomicAdd(c, sum);
//WRONG: *c += sum; This read-write operation must be atomic!
}
}
// Initialize the vectors:
void init_vector(num_t *x)
{
int i;
for (i=0 ; i<N ; i++){
x[i] = 0.001 * i;
}
}
// MAIN
int main(void)
{
num_t *a, *b, *c;
num_t *dev_a, *dev_b, *dev_c;
size_t size = N * sizeof(num_t);
cudaMalloc((void**)&dev_a, size);
cudaMalloc((void**)&dev_b, size);
cudaMalloc((void**)&dev_c, size);
a = (num_t*)malloc(size);
b = (num_t*)malloc(size);
c = (num_t*)malloc(size);
init_vector(a);
init_vector(b);
cudaMemcpy(dev_a, a, size, cudaMemcpyHostToDevice);
cudaMemcpy(dev_b, b, size, cudaMemcpyHostToDevice);
dot<<<N/THREADS_PER_BLOCK, THREADS_PER_BLOCK>>>(dev_a, dev_b, dev_c);
cudaMemcpy(c, dev_c, sizeof(num_t), cudaMemcpyDeviceToHost);
printf("a = [\n");
int i;
for (i=0;i<10;i++){
printf("%g\n",a[i]);
}
printf("...\n");
for (i=N-10;i<N;i++){
printf("%g\n",a[i]);
}
printf("]\n\n");
printf("a*b = %g.\n", *c);
free(a); free(b); free(c);
cudaFree(dev_a);
cudaFree(dev_b);
cudaFree(dev_c);
}
을 내가 그것을 컴파일 : http://pastebin.com/8yTzXUuK에서 찾을 수 있습니다 내 NVIDIA 카드에 대한
/usr/local/cuda-5.0/bin/nvcc -m64 -I/usr/local/cuda-5.0/include -gencode arch=compute_20,code=sm_20 -o multi_dot_product.o -c multi_dot_product.cu
g++ -m64 -o multi_dot_product multi_dot_product.o -L/usr/local/cuda-5.0/lib64 -lcudart
정보 (this tutorial 기준) 내 CUDA 코드는 다음과 같다.
N = 2048 * 8;
a = zeros(N,1);
for i=1:N
a(i) = 0.001*(i-1);
end
dot_product = a'*a;
그러나
가 N이 증가함에 따라, 내가 N = 2048 * 32 CUDA를 들어, 예를 들어 (상당히 다른 결과를 얻고있다 6.73066e + 07 동안 reutrns : 나는 다음과 같은 간단한 코드를 사용하여 MATLAB에서 결과를 확인하기 위해 시도 MATLAB은 9.3823e + 07을 반환하고, N = 2048 * 64의 경우 CUDA는 3.28033e + 08을 제공하는 반면 MATLAB은 7.5059e + 08을 제공합니다. 필자는 불일치가 내 C 코드에서float의 사용으로 인해 발생한다고 생각하지만,
double으로 바꾸면 컴파일러에서
atomicAdd이 이중 매개 변수를 지원하지 않는다고 불평합니다. 이 문제를 어떻게 해결해야합니까?
는 업데이트 : 또한, N (예를 들어, 2048 * 64)의 높은 값을, 나는 CUDA에 의해 반환되는 결과가 모든 실행에 변경 것으로 나타났습니다. N이 낮 으면 (예 : 2048 * 8) 이는 발생하지 않습니다.
동시에 더 기본적인 질문이 있습니다. 변수 temp은 크기가 THREADS_PER_BLOCK 인 배열이며 동일한 블록의 스레드간에 공유됩니다. 블록간에 공유되거나 모든 블록이이 변수의 다른 복사본에서 작동합니까? 모든 블록에 대한 지침으로 dot 메서드를 생각해야합니까?
// *c = 0.00;
을 그리고 커널 전에 호스트 코드에 다음 줄을 추가 : 누군가가 당신의 커널에서 변수가이 예에서 공유하는 방법을 정확하게 작업을 분할하는 방법
"상당히 다른 결과"가 무엇인지 계량 할 수 있습니까? 상대적인 오류와 절대적 오류는 무엇입니까? – talonmies
@talonmies 예, 질문을 업데이트했습니다. –
모든 블록은'temp'를 포함하여 주어진'__shared__' 변수의 다른 사본에서 작동합니다. [here] (http://docs.nvidia.com/cuda/cuda-c-programming-guide/index.html#atomic-functions)에 설명 된 방법을 사용하여'double' atomicAdd를 수행 할 수 있습니다. –