병렬 프로그래밍 for 신경망

GPU 활용

유닛은 신경망을 구축하는 최소 단위이다. 신경망에는 방대한 양의 유닛이 들어가기 때문에, 한정된 계산자원으로 효율적으로 계산하는 것은 매우 중요하다. cpu는 병렬 연산이 가능하지만 cpu의 코어 수는 한정되어 있기 때문에, gpu를 이용하여 병렬 연산을 실행할 수 있다. 

이미 구현되어 있는 행렬 연산 프레임워크를 사용하기 때문에, 구현 과정에 대한 지식 없이 이미 작성되어 있는 것을 사용할 수는 있다. 하지만 이번 장에서는 새로운 것을 만들거나, 예상치 못한 상황에 부딪혔을 때 활용가능한 응용력을 기르기 위해, 기초적인 것을 다룰 예정이다. 

CUDA 프로그래밍

CUDA kernel

CUDA kernel은 실행의 최소 단위를 정의하는 함수이다. CUDA 커널의 정의된 내용을 복수의 스레드가 동시에 실행한다. 아래는 kernel 정의에 예시이다. 

__global__ void MatAdd(float A[N][N], float B[N][N], float c[N][N])
{
    int i = threadIdx.x;
    int j = threadIdx.y;
    C[i][j] = A[i][j] + B[i][j];
 }
 
 int main()
 {
    float *A, *B, *C;
    int N = 100;
    
    cudaMalloc((void**)&A, N*N*sizeof(float));
    cudaMalloc((void**)&B, N*N*sizeof(float));
    cudaMalloc((void**)&C, N*N*sizeof(float));
    
    float *a = malloc(N*N*sizeof(float));
    float *b = malloc(N*N*sizeof(float));
    float *c = malloc(N*N*sizeof(float));
    
    cudaMemcpy (A, a, N * N * sizeof(A), cudaMemcpyHostToDevice);
    cudaMemcpy (B, b, N * N * sizeof(B), cudaMemcpyHostToDevice);
    
    int numBlocks = 1;
    dim3 threadsPerBlock(N, N);
    MatAdd<<<numBlocks, threadsPerBlock>>> (A,B,C);
    
    cudaMemcpy(c, C, N*N*sizeof(*C), cudaMemcpyHostToDevice);
    
    cudaFree(A); cudaFree(B); cudaFree(C);
    }

• kernel 함수는 식별자 __global__ 을 사용한다.
• kernel은 <<<(그리드당 블록수), (블록당 스레드 수) >>> 형식의 특수한 연산자로 호출된다. 
• cudaMalloc으로 호출된 A, B, C는 디바이스 메모리로 gpu쪽 메모리에 할당된다. 
• N*N*1은 실행 스레드 수를 나타낸다. 
• threadIdx는 CUDA의 예약 변수이고, 현재 kernel이 어떤 스레드에서 실행되는지를 나타낸다.

Thread 구성

Thread는 CUDA에서 정의되는 실행의 최소 단위로, threadIdx.x, threadIdx.y, threadIdx.z의 3차원으로 정의된다. 활용할 수 있는 최대 thread 수는 하드웨어마다 다르다. 

• Thread를 묶은 그룹을 Block이라고 부른다. Block은 3차원으로 정의되고, 포함할 수 있는 thread 수는 하드웨어마다 정해져 있다. 
• Block을 묶은 것을 Grid라고 부른다. 

CPU memory, GPU memory 

CUDA에서는 기본적으로 cpu memory와 gpu memory를 다른 메모리 공간으로 다룬다. 아래 그림은 각 메모리 공간에서의 메모리 전송 절차를 나타낸 것이다. 

cudaError_t error = cudaMemcpy(a, A, M*N*sizeof(*a), cudaMemcpyHostToDevice);

• cudaMemcpy 함수
  • 2번째 인수인 복사 원본에서 1번째 인수로 메모리 내용을 복사한다. 
  • 3번째 인수는 복사 원본에서 얼마만큼 메모리를 복사할지 선택한다.
  • cudaMemcpyHostToDevice 
    • 호스트에서 디바이스 메모리로 전송 
  • cudaMemcpyDeviceToHost 
    • 디바이스에서 호스트 메모리로 전송
  • cudaMemcpyDeviceToDevice
    • 디바이스에서 디바이스 메모리로 전송
• 메모리간 전송은 각 메모리 공간 내에서의 전송보다 상당히 큰 비용이 소모된다.
• CUDA의 디바이스 메모리는 글로벌 메모리와 공유 메모리로 나뉜다. 
  • 공유메모리는 각 스레드에서 공유할 수 있는 빠른 메모리로, 글로벌 메모리보다 용량이 적다.