C++ CUDA 문제 2. CPU에서 GPU로 데이터를 전달하여 3차원 공간에서 병렬 실행하기

주의할 점은 cudaMalloc을 사용할 시, 함수 인자로 이중포인터(void**(&d_data))를 넘겨야 한다.

사람들이 왜 cudaMalloc을 사용할 시 이중포인터를 함수 인자로 넣어야 하는지 궁금해 하는 경우가 많다.

 

간단하게 이유를 설명 하자면 d_data의 주소(&d_data)는 cpu main 함수 스택에 저장 된다.

cudaMalloc((void**)&d_data, array_byte_size) 에서는 gpu 내에 d_data 값에 동적 메모리 할당을 하고, 

이후 gpu 내에서 d_data 값이 새롭게 지정되며, 이 메모리에서 gpu 연산이 되기 때문이다.

그렇지만 d_data의 주소(&d_data)는 cpu main 함수 스택에서 처음과 같이 유지된다.

따라서 이중 포인터를 사용하면, cpu와 gpu에서의 브릿지 역할을 가능하게 할 수 있다.

 

#include "cuda_runtime.h"
#include "device_launch_parameters.h"
 
#include <stdio.h>
#include <stdlib.h>
#include <cstring>
#include <time.h>
 
 
__global__ void unique_idx_calc_threadIdx(int* input) {
    int tid = threadIdx.x;
    printf("threadIdx : %d, value : %d \n", tid, input[tid]);
}
 
 
 
__global__ void unique_gid_calculation_3d(int* data)
{
    int tid = threadIdx.x + threadIdx.y * blockDim.x + threadIdx.z * blockDim.y * blockDim.x;
    int x_offset = blockIdx.x * blockDim.x * blockDim.y * blockDim.z;
    int y_offset = blockIdx.y * gridDim.x * blockDim.x * blockDim.y * blockDim.z;
    int z_offset = blockIdx.z * gridDim.x * gridDim.y * blockDim.x * blockDim.y * blockDim.z;
    int gid = tid + x_offset + y_offset + z_offset;
 
    printf("blockIdx.x : %d, blockIdx.y : %d, blockIdx.z : %d, threadIdx.x : %d, gid : %d, value : %d\n",
        blockIdx.x, blockIdx.y, blockIdx.z, threadIdx.x, gid, data[gid]);
}
 
int main()
{
    int array_size = 512;
    int array_byte_size = sizeof(int) * array_size;
    time_t t;
    srand((unsigned int)time(&t));
    int* h_data = (int*)malloc(array_byte_size);
    for (int i = 0;i < array_size; i++)
    {
        int num = rand() & 0xff; // 0에서 255까지의 random value를 배열 안에 넣는다.
        printf("num value : %d\n", num);
        h_data[i] = num;
    }
 
    printf("\n \n");
 
    int* d_data;
    cudaMalloc((void**)&d_data, array_byte_size);
   cudaMemcpy(d_data, h_data, array_byte_size, cudaMemcpyHostToDevice); // Cpu에서 Device(GPU)로 데이터를 보낸다.
 
    dim3 block(2,2,2);
    dim3 grid(4,4,4);
 
    unique_gid_calculation_3d <<<grid, block >>>(d_data);
    cudaDeviceSynchronize();
    cudaDeviceReset();
    cudaFree(d_data);
    free(h_data);
    printf("Success");
 
    return 0;
}