기록용으로 적어 놓습니다.
CUDA를 공부(전 https://class.coursera.org/hetero-002 여기에서 공부하고 있습니다.)하다보니, device 메모리에 값 넣어놓고 cudafree 쓰기 전까지 메모리에 남아 있는 값을 재활용할 수 있는지, 그리고 거기에 새로운 계산을 할 수 있는지 궁금해지더군요.
그래서 간단한 더하기 프로그래밍 예제를 조금 수정해보았습니다.
(참조 : http://developer.download.nvidia.com/books/cuda-by-example/cuda-by-example-sample.pdf)
---------------------------------
#include <stdio.h>
#define N 10
static void HandleError( cudaError_t err,
const char *file,
int line ) {
if (err != cudaSuccess) {
printf( "%s in %s at line %d\n", cudaGetErrorString( err ),
file, line );
}
}
#define HANDLE_ERROR( err ) (HandleError( err, __FILE__, __LINE__ ))
__global__ void minusone(int *c)
{
unsigned int t = threadIdx.x;
unsigned int tx = blockIdx.x * blockDim.x + t;
if (tx < N)
c[tx] -= 1;
}
__global__ void add( int *a, int *b, int *c )
{
int tid = blockIdx.x; // handle the data at this index
if (tid < N)
c[tid] = a[tid] + b[tid];
}
int main(void)
{
int a[N], b[N], c[N];
int *dev_a, *dev_b, *dev_c;
HANDLE_ERROR( cudaMalloc( (void**)&dev_a, N * sizeof(int) ) );
HANDLE_ERROR( cudaMalloc( (void**)&dev_b, N * sizeof(int) ) );
HANDLE_ERROR( cudaMalloc( (void**)&dev_c, N * sizeof(int) ) );
for (int i=0; i<N; i++)
{
a[i] = -i;
b[i] = i * i;
}
HANDLE_ERROR( cudaMemcpy( dev_a, a, N * sizeof(int), cudaMemcpyHostToDevice ) );
HANDLE_ERROR( cudaMemcpy( dev_b, b, N * sizeof(int), cudaMemcpyHostToDevice ) );
add<<<N,1>>>( dev_a, dev_b, dev_c );
HANDLE_ERROR( cudaMemcpy( c, dev_c, N * sizeof(int), cudaMemcpyDeviceToHost ) );
for (int i=0; i<N; i++)
{
printf( "%d + %d = %d\n", a[i], b[i], c[i] );
}
minusone<<<N,1>>>( dev_c );
HANDLE_ERROR( cudaMemcpy( c, dev_c, N * sizeof(int), cudaMemcpyDeviceToHost ) );
for (int i=0; i<N; i++)
{
printf( "%d + %d - 1 = %d\n", a[i], b[i], c[i] );
}
cudaFree( dev_a );
cudaFree( dev_b );
cudaFree( dev_c );
for (int i=0; i<N; i++)
{
printf( "%d + %d - 1 = %d\n", a[i], b[i], c[i] );
}
return 0;
}
------------------------
참조의 42쪽에 있는 코드에다가 굵은 글자만 넣어봤습니다. block과 dim 사이즈를 같이 하고, 앞의 cuda kernel에서 실행한 결과가 들어있는 메모리인 dev_c를 그대로 새로운 cuda kernel에 넣고, 새로운 cuda kernel에서는 각 값에 1을 빼줍니다.
그랬더니 되더군요.
kernel 함수 안 에서의 코딩 개념이 CPU랑 달라서 헷갈리는데, 그걸 이런 식으로 쪼개면 이해하기 쉽지 않을까 싶네요.
댓글 없음:
댓글 쓰기