3D Client Programmer

10개의 원소를 갖는 두 벡터 a, b의 합을 그림으로 보면 다음과 같다.

그림 - 벡터의 합

이것을 우리가 흔히 사용했던 C 코드로 작성을 하면 다음과 같을 것이다.

그림 - 벡터의 합 C 코드

그림 - 벡터의 합 결과

이것을 이제 CUDA를 이용하여 병렬 프로그래밍을 해보도록 하자.

그림 - CUDA를 이용한 벡터의 합 Main

이전 C코드에서 Main부분은 바뀐 부분이 GPU에서 사용되는 변수가 추가된 것과 add() 함수가 커널 함수로 바뀐 것 밖에 없다.

커널 함수 add<<<N, 1>>>의 첫 번째 인자를 통해 N개의 병렬 블록(Block)들이 실행되도록 지시하고 뒤에 1은 스레드가 1개가 돈다는 뜻이다. 즉 add<<<Block 개수, Thread 개수>>> 이다. 

Block과 Thread는 다음 글에서 다시 설명하도록 하겠다.

이 블록(Block)이라는 것은 쉽게 말해 각각의 병렬 실행을 뜻한다. 즉 N개의 add()가 생성되어 병렬로 실행된다는 것이다.

블록의 개수 N은 65,535의 값을 초과할 수 없다.(하드웨어의 제약사항)

그림 - add 커널 함수

add() 함수는 tid 값에 처음 보는 blockIdx.x가 대입되었다. 앞에서 커널 함수 add()를 호출할 때 블록 인자를 넣어주었다.

이 blockIdx가 바로 디바이스 코드를 작동하는 블록의 인덱스 값이다. 그리고 'x, y'가 있지만, 여기서는 'x'만 이용하였을 뿐이다.(y는 어차피 1이니...)

그리고 밑에서 while 문이 if 문으로 바뀌었다. 이것은 블록별로 병렬 실행되므로 while로 계산할 필요없이, 해당하는 인덱스의 데이터만 처리하면 되기 때문이다.

각 스레드들의 blockIdx.x는 0에서 N - 1사이의 값을 가질 것이다. 따라서 변수 blockIdx.x가 서로 다른 값을 가지면서 동일한 디바이스 코드의 복사본들을 수행하는 네 개의 블록들을 가정해볼 수 있다.

그림 - GPU를 이용한 벡터의 합

그리고 이러한 병렬 블록들의 집합을 그리드(grid)라고 부른다. 

CUDA에서는 CPU를 호스트, GPU는 디바이스라고 부른다.

그리고 호스트 코드는 CPU에서 수행될 코드를 의미하고, 디바이스 코드는 CUDA가 가능한 그래픽 칩셋에서 수행될 코드를 의미한다.

__global__ : 커널(Kernel) 함수임을 컴파일러에게 명시

__device__ :  디바이스(GPU)에서만 사용하는 함수라고 명시 (생략이 가능함)

__host__ : 호스트(CPU)에서 실행되는 함수라고 명시

커널(Kernel) 함수 : 디바이스에서 실행되는 함수이며, 호스트에서만 호출 됨

<<<A,B>>> : CUDA 런타임 시스템에 넘겨질 매개변수

아래 예제는 단순하게 A + B = C 예제이다.

cudaMalloc() : 디바이스 메모리(전역 메모리)를 할당

cudaMemcpy() : 호스트와 디바이스간의 메모리 복사

 - cudaMemcpyDeviceToHost

 - cudaMemcpyHostToDevice

 - cudaMemcpyDeviceToDevice

cudaFree() : 디바이스 메모리 해제

* cudaMalloc()으로 할당한 메모리 포인터를 커널로 전달할 수 있다.

* 커널에서 cudaMalloc()으로 할당한 메모리 포인터를 이용하여 메모리를 읽거나 쓸 수 있다.

* cudaMalloc()으로 할당한 메모리 포인터를 호스트 함수로 전달할 수 있다.

* 호스트 함수에서 cudaMalloc()으로 할당한 메모리 포인터를 읽거나 쓸 수 없다.