적절한 block_size 의 지정 - GPU 정보의 확인

GPU 정보 확인

nvidia-smi 명령어를 통한 확ㅇ

+-----------------------------------------------------------------------------------------+
| NVIDIA-SMI 561.17                 Driver Version: 561.17         CUDA Version: 12.6     |
|-----------------------------------------+------------------------+----------------------+
| GPU  Name                  Driver-Model | Bus-Id          Disp.A | Volatile Uncorr. ECC |
| Fan  Temp   Perf          Pwr:Usage/Cap |           Memory-Usage | GPU-Util  Compute M. |
|                                         |                        |              
 MIG M. |
|=========================================+========================+======================|
|   0  NVIDIA GeForce RTX 3080 ...  WDDM  |   00000000:01:00.0  On |              
    N/A |
| N/A   51C    P5             22W /   96W |     704MiB /  16384MiB |      1%      Default |
|                                         |                        |              
    N/A |
+-----------------------------------------+------------------------+----------------------+
                                                                                  

+-----------------------------------------------------------------------------------------+
| Processes:                                                                      
        |
|  GPU   GI   CI        PID   Type   Process name                              GPU Memory |
|        ID   ID                                                               Usage      |
|=========================================================================================|
|    0   N/A  N/A      7208    C+G   ...Programs\Microsoft VS Code\Code.exe      N/A      |
|    0   N/A  N/A     12248    C+G   ...ekyb3d8bbwe\PhoneExperienceHost.exe      N/A      |
|    0   N/A  N/A     12688    C+G   ...siveControlPanel\SystemSettings.exe      N/A      |
|    0   N/A  N/A     15632    C+G   ...2txyewy\StartMenuExperienceHost.exe      N/A      |
|    0   N/A  N/A     17132    C+G   ...n\131.0.2903.112\msedgewebview2.exe      N/A      |
|    0   N/A  N/A     23192    C+G   ...cal\Microsoft\OneDrive\OneDrive.exe      N/A      |
|    0   N/A  N/A     24204    C+G   ...oogle\Chrome\Application\chrome.exe      N/A      |
/A      |
|    0   N/A  N/A     26280    C+G   ...n\131.0.2903.112\msedgewebview2.exe      N/A      |
|    0   N/A  N/A     31092    C+G   ...5n1h2txyewy\ShellExperienceHost.exe      N/A      |
+-----------------------------------------------------------------------------------------+

 

CUDA 프로그램에서의 GPU의 한계값과 성능의 직접 확인

#include <iostream>
#include <cuda_runtime.h>

int main() {
    cudaDeviceProp prop;
    cudaGetDeviceProperties(&prop, 0); // 0번 GPU 사용

    std::cout << "Device Name: " << prop.name << "\n";
    std::cout << "Max Threads Per Block: " << prop.maxThreadsPerBlock << "\n";
    std::cout << "Max Threads Dim: (" 
              << prop.maxThreadsDim[0] << ", " 
              << prop.maxThreadsDim[1] << ", " 
              << prop.maxThreadsDim[2] << ")\n";
    std::cout << "Max Grid Size: (" 
              << prop.maxGridSize[0] << ", " 
              << prop.maxGridSize[1] << ", " 
              << prop.maxGridSize[2] << ")\n";
    return 0;
}

PS C:\Users\owner\Desktop\AI_framework-dev\tests\GPU_test> .\GPU.exe       
Device Name: NVIDIA GeForce RTX 3080 Ti Laptop GPU
Max Threads Per Block: 1024
Max Threads Dim: (1024, 1024, 64)
Max Grid Size: (2147483647, 65535, 65535)

블록 차원의 최대 크기 1024, 1024, 64 각 블록에서의 최대 스레드 차원, 실제론 스레드 수의 총합이 1024 를 초과하지 않도록 제한해야 한다.

그리드 차원의 최대 크기

그리드 차원은 총 작업량에 따라 동적으로 설정

데이터가 1,000,000 개의 요소로 구성된 경우

int block_size = 256;
int grid_size = (1000000 + block_size - 1) / block_size; // 올림 계산
kernel<<<grid_size, block_size>>>(...);

 

스레드 차원

2D, 3D 데이터 처리 시 블록 차원을 활용

1024, 1024 행렬의 경우

dim3 threadsPerBlock(32, 32); // 블록당 32x32 스레드
dim3 blocksPerGrid((N + 31) / 32, (M + 31) / 32); // 그리드 크기
kernel<<<blocksPerGrid, threadsPerBlock>>>(...);

 

 

#include <stdio.h>
#include <cuda_runtime.h>

#define N 1024 * 1024 // 데이터 크기

__global__ void kernel(float *data) {
    int idx = threadIdx.x + blockIdx.x * blockDim.x;
    if (idx < N) {
        data[idx] += 1.0f; // 간단한 작업
    }
}

int main() {
    float *d_data;
    cudaMalloc(&d_data, N * sizeof(float));

    int blockSizes[] = {32, 64, 128, 256, 512, 1024};
    for (int blockSize : blockSizes) {
        int gridSize = (N + blockSize - 1) / blockSize;

        cudaEvent_t start, stop;
        cudaEventCreate(&start);
        cudaEventCreate(&stop);

        cudaEventRecord(start);
        kernel<<<gridSize, blockSize>>>(d_data);
        cudaEventRecord(stop);

        cudaEventSynchronize(stop);
        float milliseconds = 0;
        cudaEventElapsedTime(&milliseconds, start, stop);
        printf("Block Size: %d, Time: %.3f ms\n", blockSize, milliseconds);

        cudaEventDestroy(start);
        cudaEventDestroy(stop);
    }

    cudaFree(d_data);
    return 0;
}

컴파일하여 exe 파일 생성

nvcc -arch=sm_80 -o block_size.exe block_size.cu

실행 결과 확인

PS C:\Users\owner\Desktop\AI_framework-dev\tests\GPU_test> .\block_size.exe
Block Size: 32, Time: 0.588 ms
Block Size: 64, Time: 0.040 ms
Block Size: 128, Time: 0.025 ms
Block Size: 256, Time: 0.024 ms
Block Size: 512, Time: 0.070 ms
Block Size: 1024, Time: 0.042 ms

256의 블록 크기에서 가장 낮은 실행 시간이 나왔음, 이는 해당 GPU 에서 256의 블록 크기가 가장 효율적인 것으로 확인할 수 있다.