VSCode 에서 CUDA (nvcc) 개발 환경 설정하기 – 1

이전 포스팅에서 그래픽 카드 붙은 시스템 까지는 구성해놨으니 이제 개발환경을 설정해본다. 사실 NVIDIA Container Runtime 만 구성하면 NVIDIA NGC 에서 CUDA 뿐만 아니라 여러 라이브러리, 프레임워크 버전별 이미지를 제공해주다보니 직접 버전별로 구성해놓을 필요가 없어서 시스템을 깔끔하게 관리할 수도 있고 집에 있는 개인 장비, Cloud Workspace 그리고 폐쇄망인 회사 업무 환경에 까지 일관된 개발환경을 구성할 수 있을 것 같다.

VSCode 에서 Docker 와 Dev Conatainers 확장만 이용해도 제법 그럴싸할 것 같은데 일단 NGC 에서 아무 개발 이미지를 하나 대충 받아다가

vim docker-compose.yml

version: '3'

services:
  nvcc:
    image: nvcr.io/nvidia/cuda:11.3.0-devel-ubuntu20.04
    command: nvcc matmul.cu
    working_dir: /src
    volumes:
      - ./src:/src

무려 8년 전, 학부시절 만들어놓은 matmul 코드를 돌려본다.

vim src/matmul.cu

#include<memory>
#include<cuda.h>

__global__ void MatrixMulKernel(float* Md, float* Nd, float* Pd, int width){
    int tx = threadIdx.x;
    int ty = threadIdx.y;
    float Pvalue = 0;
    for (int k = 0; k < width; ++k) {
        float Mdelement = Md[ty * width + k];
        float Ndelement = Nd[k * width + tx];
        Pvalue += Mdelement * Ndelement;
    }

    Pd[ty * width + tx]=Pvalue;
}

void MatrixMultiplication(float* M, float* N, float* P, int width) {
    int size = width * width * sizeof(float);
    float *Md, *Nd, *Pd;

    cudaMalloc((void**) &Md, size);
    cudaMemcpy(Md, M, size, cudaMemcpyHostToDevice);
    cudaMalloc((void**) &Nd, size);
    cudaMemcpy(Nd, N, size, cudaMemcpyHostToDevice);

    cudaMalloc((void**) &Pd, size);

    dim3 dimBlock(width, width);
    dim3 dimGrid(1, 1);

    MatrixMulKernel<<<dimGrid, dimBlock>>>(Md, Nd, Pd, width);

    cudaMemcpy(P, Pd, size, cudaMemcpyDeviceToHost);
    cudaFree(Md);
    cudaFree(Nd);
    cudaFree(Pd);
}

int main(){
    clock_t start;
    clock_t finish;
    double duration;
    start = clock();
    printf("program start\n");

    int width = 1000;   
    float *M, *N, *P;
    M = (float *)malloc(width * width * sizeof(float));
    N = (float *)malloc(width * width * sizeof(float));
    P = (float *)malloc(width * width * sizeof(float));

    for (int i = 0; i < width * width; i++) {
        if(i % 2) M[i] = 0;
        else M[i] = 1;
    }
    for (int i = 0; i < width * width; i++) {
        if(i % 2) N[i] = 1;
        else N[i] = 0;
    }

    MatrixMultiplication(M, N, P, width);
    finish = clock();
    printf("program finished\n");

    duration = (double)(finish - start) / CLOCKS_PER_SEC;
    printf("%f sec\n", duration);

    return 0;
}

이 프로젝트랑 같이 있던 Makefile 에서, 당시 최신 버전이었을, cuda-7.0 쓰도록 명시해놨길래 분명 컴파일이 안 될 줄 알았는데 너무 기본 api 만 사용하는 프로젝트여서 그런지 신기하게도 잘 돌아간다.

# docker compose up
[+] Running 1/0
 ⠿ Container matmul-nvcc-1  Created                                                                           0.0s
...
matmul-nvcc-1 exited with code 0
# src/a.out 
program start
program finished
0.095565 sec

이어서 VSCode 가 C/C++ IDE 로 얼마나 쓸만할지 이것저것 확장들을 조합해 개발 환경 먼저 구축해볼텐데 거기에서 파이썬과 상호작용할 수 있는 CUDA 모듈까지 간단하게 만들어보면 재미있을 것 같다.