cuda 튜토리얼
cuda 시작하기

비고

버전

선결 요건

CUDA로 프로그래밍을 시작하려면 CUDA 툴킷과 개발자 드라이버를 다운로드하여 설치하십시오. 이 툴킷에는 nvcc , NVIDIA CUDA Compiler 및 CUDA 응용 프로그램 개발에 필요한 기타 소프트웨어가 포함되어 있습니다. 드라이버는 GPU 프로그램이 CUDA 가능 하드웨어 에서 올바르게 실행되도록 보장합니다.

명령 줄에서 nvcc --version 을 실행하여 CUDA 툴킷이 컴퓨터에 올바르게 설치되었는지 확인할 수 있습니다. 예를 들어, 리눅스 머신에서,

$ nvcc --version
nvcc: NVIDIA (R) Cuda compiler driver
Copyright (c) 2005-2016 NVIDIA Corporation
Built on Tue_Jul_12_18:28:38_CDT_2016
Cuda compilation tools, release 8.0, V8.0.32

컴파일러 정보를 출력합니다. 이전 명령이 성공적이지 않은 경우 CUDA 툴킷이 설치되지 않았거나 nvcc (Windows 시스템의 경우 C:\CUDA\bin , POSIX OS의 경우 /usr/local/cuda/bin ) 경로는 사용자의 일부가 아닙니다 PATH 환경 변수.

또한 nvcc 를 사용하여 CUDA 프로그램을 컴파일하고 빌드하는 호스트 컴파일러가 필요합니다. Windows의 경우 Microsoft Visual Studio와 함께 제공되는 Microsoft 컴파일러 인 cl.exe 입니다. POSIX OS에서는 gcc 나 g++ 포함한 다른 컴파일러를 사용할 수 있습니다. 공식 CUDA Quick Start Guide 는 특정 플랫폼에서 지원되는 컴파일러 버전을 알려줍니다.

모든 것이 올바르게 설정되었는지 확인하려면, 모든 도구가 올바르게 작동하는지 확인하기 위해 간단한 CUDA 프로그램을 컴파일하고 실행합시다.

__global__ void foo() {}

int main()
{
  foo<<<1,1>>>();

  cudaDeviceSynchronize();
  printf("CUDA error: %s\n", cudaGetErrorString(cudaGetLastError()));

  return 0;
}

이 프로그램을 컴파일하려면 test.cu 파일에 복사하고 명령 행에서 컴파일하십시오. 예를 들어, Linux 시스템에서는 다음이 작동합니다.

$ nvcc test.cu -o test
$ ./test
CUDA error: no error

프로그램이 오류없이 성공하면 코딩을 시작하십시오!

CUDA로 두 배열 합계

이 예제는 두 개의 int 배열을 CUDA로 더하는 간단한 프로그램을 만드는 법을 보여줍니다.

CUDA 프로그램은 이기종이며 CPU와 GPU 모두에서 실행됩니다.

CUDA를 사용하는 프로그램의 주요 부분은 CPU 프로그램과 유사하며

• GPU에서 사용될 데이터에 대한 메모리 할당
• 호스트 메모리에서 GPU 메모리로 데이터 복사
• 커널 함수를 호출하여 데이터 처리하기
• 결과를 CPU 메모리로 복사

장치 메모리를 할당하기 위해 우리는 cudaMalloc 함수를 사용합니다. 장치와 호스트간에 데이터를 복사하려면 cudaMemcpy 기능을 사용할 수 있습니다. cudaMemcpy 의 마지막 인수는 복사 작업의 방향을 지정합니다. 가능한 유형은 5 가지입니다.

• cudaMemcpyHostToHost - 호스트 -> 호스트
• cudaMemcpyHostToDevice - 호스트 -> 장치
• cudaMemcpyDeviceToHost - 기기 -> 호스트
• cudaMemcpyDeviceToDevice - 장치 -> 장치
• cudaMemcpyDefault - 기본 기반 통합 가상 주소 공간

다음으로 커널 함수가 호출됩니다. 트리플 셰브론 사이의 정보는 실행 구성이며 커널을 병렬로 실행하는 장치 스레드의 수를 나타냅니다. 첫 번째 숫자 (예 : 2 )는 블록 수와 두 번째 숫자 (예 : (size + 1) / 2 ) - 블록의 스레드 수를 지정합니다. 이 예제에서는 크기에 1을 더하여 하나의 스레드가 두 개의 요소를 담당하는 대신 하나의 추가 스레드를 요청합니다.

커널 호출은 비동기 함수이기 때문에 실행이 완료 될 때까지 대기하도록 cudaDeviceSynchronize 가 호출됩니다. 결과 배열은 호스트 메모리에 복사되고 장치에 할당 된 모든 메모리는 cudaFree 로 해제됩니다.

함수를 커널로 정의하려면 __global__ 선언 지정자가 사용됩니다. 이 함수는 각 스레드에 의해 호출됩니다. 각 스레드가 결과 배열의 요소를 처리하도록하려면 각 스레드를 구별하고 식별하는 방법이 필요합니다. CUDA는 blockDim , blockIdx 및 threadIdx 변수를 정의합니다. 미리 정의 된 변수 인 blockDim 에는 커널 실행을위한 두 번째 실행 구성 매개 변수에 지정된대로 각 스레드 블록의 크기가 포함됩니다. 미리 정의 된 변수 인 threadIdx 및 blockIdx 에는 각각 해당 스레드 블록 내의 스레드 및 격자 내의 스레드 블록의 인덱스가 포함됩니다. 잠재적으로 배열의 요소보다 하나 이상의 스레드를 요청할 것이므로 어레이의 끝을 지나서 액세스하지 못하도록 size 를 전달해야합니다.

#include "cuda_runtime.h"
#include "device_launch_parameters.h"

#include <stdio.h>

__global__ void addKernel(int* c, const int* a, const int* b, int size) {
    int i = blockIdx.x * blockDim.x + threadIdx.x;
    if (i < size) {
        c[i] = a[i] + b[i];
    }
}

// Helper function for using CUDA to add vectors in parallel.
void addWithCuda(int* c, const int* a, const int* b, int size) {
    int* dev_a = nullptr;
    int* dev_b = nullptr;
    int* dev_c = nullptr;

    // Allocate GPU buffers for three vectors (two input, one output)
    cudaMalloc((void**)&dev_c, size * sizeof(int));
    cudaMalloc((void**)&dev_a, size * sizeof(int));
    cudaMalloc((void**)&dev_b, size * sizeof(int));

    // Copy input vectors from host memory to GPU buffers.
    cudaMemcpy(dev_a, a, size * sizeof(int), cudaMemcpyHostToDevice);
    cudaMemcpy(dev_b, b, size * sizeof(int), cudaMemcpyHostToDevice);

    // Launch a kernel on the GPU with one thread for each element.
    // 2 is number of computational blocks and (size + 1) / 2 is a number of threads in a block
    addKernel<<<2, (size + 1) / 2>>>(dev_c, dev_a, dev_b, size);
    
    // cudaDeviceSynchronize waits for the kernel to finish, and returns
    // any errors encountered during the launch.
    cudaDeviceSynchronize();

    // Copy output vector from GPU buffer to host memory.
    cudaMemcpy(c, dev_c, size * sizeof(int), cudaMemcpyDeviceToHost);

    cudaFree(dev_c);
    cudaFree(dev_a);
    cudaFree(dev_b);
}

int main(int argc, char** argv) {
    const int arraySize = 5;
    const int a[arraySize] = {  1,  2,  3,  4,  5 };
    const int b[arraySize] = { 10, 20, 30, 40, 50 };
    int c[arraySize] = { 0 };

    addWithCuda(c, a, b, arraySize);

    printf("{1, 2, 3, 4, 5} + {10, 20, 30, 40, 50} = {%d, %d, %d, %d, %d}\n", c[0], c[1], c[2], c[3], c[4]);

    cudaDeviceReset();

    return 0;
}

CUDA 스레드를 하나 시작하여 인사 해보자.

이 간단한 CUDA 프로그램은 GPU (일명 "장치")에서 실행할 함수를 작성하는 방법을 보여줍니다. CPU 또는 "호스트"는 "커널"이라는 특수 함수를 호출하여 CUDA 스레드를 생성합니다. CUDA 프로그램은 추가 구문이있는 C ++ 프로그램입니다.

어떻게 작동하는지 보려면 hello.cu 파일에 다음 코드를 입력하십시오.

#include <stdio.h>

// __global__ functions, or "kernels", execute on the device
__global__ void hello_kernel(void)
{
  printf("Hello, world from the device!\n");
}

int main(void)
{
  // greet from the host
  printf("Hello, world from the host!\n");

  // launch a kernel with a single thread to greet from the device
  hello_kernel<<<1,1>>>();

  // wait for the device to finish so that we see the message
  cudaDeviceSynchronize();

  return 0;
}

(장치에서 printf 함수를 사용하려면 최소한 2.0의 계산 기능을 갖춘 장치가 필요합니다. 자세한 내용은 버전 개요 를 참조하십시오.)

이제 NVIDIA 컴파일러를 사용하여 프로그램을 컴파일하고 실행 해 봅시다.

$ nvcc hello.cu -o hello
$ ./hello
Hello, world from the host!
Hello, world from the device!

위의 예에 대한 몇 가지 추가 정보 :

• nvcc 는 "NVIDIA CUDA Compiler"의 약자입니다. 소스 코드를 호스트 및 장치 구성 요소로 분리합니다.
• __global__ 은 함수 선언에 사용 된 CUDA 키워드로, 함수가 GPU 장치에서 실행되고 호스트에서 호출됨을 나타냅니다.
• 삼각 괄호 ( <<< , >>> )는 호스트 코드에서 장치 코드 ( "커널 시작"이라고도 함)로 호출을 표시합니다. 이 괄호 안의 숫자는 병렬로 실행될 횟수와 스레드 수를 나타냅니다.

샘플 프로그램 컴파일 및 실행

NVIDIA 설치 가이드는 CUDA 툴킷의 설치를 확인하기 위해 샘플 프로그램을 실행하는 것으로 끝나지 만, 어떻게 명시 적으로 언급하지는 않습니다. 먼저 모든 전제 조건을 확인하십시오. 샘플 프로그램의 기본 CUDA 디렉토리를 확인하십시오. 존재하지 않는다면 공식 CUDA 웹 사이트에서 다운로드 할 수 있습니다. 예제가있는 디렉토리로 이동하십시오.

$ cd /path/to/samples/
$ ls

다음과 비슷한 결과가 나타납니다.

0_Simple     2_Graphics  4_Finance      6_Advanced       bin     EULA.txt
1_Utilities  3_Imaging   5_Simulations  7_CUDALibraries  common  Makefile

Makefile 이이 디렉토리에 있는지 확인하십시오. UNIX 기반 시스템의 make 명령은 모든 샘플 프로그램을 빌드합니다. 또는 다른 Makefile 이있는 하위 디렉토리로 이동하여 거기에서 make 명령을 실행하여 해당 샘플 만 빌드하십시오.

두 개의 제안 된 샘플 프로그램 ( deviceQuery 및 bandwidthTest 실행하십시오.

$ cd 1_Utilities/deviceQuery/
$ ./deviceQuery 

결과물은 아래와 같습니다.

./deviceQuery Starting...

 CUDA Device Query (Runtime API) version (CUDART static linking)

Detected 1 CUDA Capable device(s)

Device 0: "GeForce GTX 950M"
  CUDA Driver Version / Runtime Version          7.5 / 7.5
  CUDA Capability Major/Minor version number:    5.0
  Total amount of global memory:                 4096 MBytes (4294836224 bytes)
  ( 5) Multiprocessors, (128) CUDA Cores/MP:     640 CUDA Cores
  GPU Max Clock rate:                            1124 MHz (1.12 GHz)
  Memory Clock rate:                             900 Mhz
  Memory Bus Width:                              128-bit
  L2 Cache Size:                                 2097152 bytes
  Maximum Texture Dimension Size (x,y,z)         1D=(65536), 2D=(65536, 65536), 3D=(4096, 4096, 4096)
  Maximum Layered 1D Texture Size, (num) layers  1D=(16384), 2048 layers
  Maximum Layered 2D Texture Size, (num) layers  2D=(16384, 16384), 2048 layers
  Total amount of constant memory:               65536 bytes
  Total amount of shared memory per block:       49152 bytes
  Total number of registers available per block: 65536
  Warp size:                                     32
  Maximum number of threads per multiprocessor:  2048
  Maximum number of threads per block:           1024
  Max dimension size of a thread block (x,y,z): (1024, 1024, 64)
  Max dimension size of a grid size    (x,y,z): (2147483647, 65535, 65535)
  Maximum memory pitch:                          2147483647 bytes
  Texture alignment:                             512 bytes
  Concurrent copy and kernel execution:          Yes with 1 copy engine(s)
  Run time limit on kernels:                     Yes
  Integrated GPU sharing Host Memory:            No
  Support host page-locked memory mapping:       Yes
  Alignment requirement for Surfaces:            Yes
  Device has ECC support:                        Disabled
  Device supports Unified Addressing (UVA):      Yes
  Device PCI Domain ID / Bus ID / location ID:   0 / 1 / 0
  Compute Mode:
     < Default (multiple host threads can use ::cudaSetDevice() with device simultaneously) >

deviceQuery, CUDA Driver = CUDART, CUDA Driver Version = 7.5, CUDA Runtime Version = 7.5, NumDevs = 1, Device0 = GeForce GTX 950M
Result = PASS

최종 Result = PASS 는 모든 것이 올바르게 작동 함을 나타냅니다. 이제 비슷한 방법으로 다른 권장 샘플 프로그램 인 bandwidthTest 를 실행하십시오. 출력은 다음과 유사합니다.

[CUDA Bandwidth Test] - Starting...
Running on...

 Device 0: GeForce GTX 950M
 Quick Mode

 Host to Device Bandwidth, 1 Device(s)
 PINNED Memory Transfers
   Transfer Size (Bytes)    Bandwidth(MB/s)
   33554432            10604.5

 Device to Host Bandwidth, 1 Device(s)
 PINNED Memory Transfers
   Transfer Size (Bytes)    Bandwidth(MB/s)
   33554432            10202.0

 Device to Device Bandwidth, 1 Device(s)
 PINNED Memory Transfers
   Transfer Size (Bytes)    Bandwidth(MB/s)
   33554432            23389.7

Result = PASS

NOTE: The CUDA Samples are not meant for performance measurements. Results may vary when GPU Boost is enabled.

다시, Result = PASS 문은 모든 것이 제대로 실행되었음을 나타냅니다. 다른 모든 샘플 프로그램은 유사한 방식으로 실행될 수 있습니다.