[OpenCL] OpenCL 정의와 동작방식(feat. Host program 작성순서)

OpenCL이란??

Open Coumputing Language의 약자로 개방형 병렬컴퓨팅 프레임워크이다. 다시말해, 이종플랫폼에서 동작하는 프로그램을 작성하기 위한 프레임워크이다. 여기서 이종플랫폼이란 CPU, GPU, FPGA 등과 같은 프로세서를 의미한다.

OpenCL은 범용 프로세서들에 대해서 쓴다면 CUDA는 GPU에 특화된 프레임 워크라고 할수있다. OpenCL을 잘 배우면 CUDA도 쉽게 배울 수 있다.

 

OpenCL 프로그램 개발

• 먼저 OpenCL 프레임워크를 사용하기 위해서는 OpenCL을 다운받고 헤더파일을 include 하여 사용할 수 있다.
• OpenCL로 프로그램을 작성할 때는 디바이스(GPU,CPU 등의 프로세서)에서 동작하는 커널 프로그램과 호스트 프로그램을 따로 작성한다.
• 커널 프로그램은 OpenCL C언어로 작성해야한다. OpcnCL C언어는 C언어와 문법이 비슷하지만 다른 언어이다.
• 호스트 프로그램은 C언어로 작성하되 OpenCL runtime API를 사용해 동작한다.

커널 프로그램은 타겟 디바이스에서 실행할 함수들을 작성하는 프로그램이고 호스트 프로그램은 타겟 디바이스에서 커널 프로그램을 실행할 수 있도록 도와주는 프로그램이다. 

 

OpenCL 동작방식

호스트 프로그램은 우리가 타겟 디바이스에서 커널 프로그램을 잘 수행할 수 있도록 도와준다. 커널 프로그램이 잘 동작하도록 하려면 호스트프로그램을 작성할 때 다음과 같은 순서대로 코드를 작성해야한다. 

 

다음예제 코드들은 https://www.eriksmistad.no/getting-started-with-opencl-and-gpu-computing/ 다음 링크에서 가져온 링크이며 덧셈 연산을 수행하는 코드이다. 

 

1. 커널 프로그램 파일 읽기

FILE *fp;
char *source_str;
size_t source_size;

fp = fopen("vector_add_kernel.cl", "r");
if (!fp) {
  fprintf(stderr, "Failed to load kernel.\n");
  exit(1);
}
source_str = (char*)malloc(MAX_SOURCE_SIZE);
source_size = fread( source_str, 1, MAX_SOURCE_SIZE, fp);
fclose( fp );

2. 플랫폼 지정과 디바이스 선택

여기서 플랫폼의 의미를 정확히 알기는 어려웠지만 터미널 창을 열어 clinfo를 했을 때 나오는 플랫폼을 의미한다. 

// Get platform and device information
cl_platform_id platform_id = NULL;
cl_device_id device_id = NULL;   
cl_uint ret_num_devices;
cl_uint ret_num_platforms;
cl_int ret = clGetPlatformIDs(1, &platform_id, &ret_num_platforms);
ret = clGetDeviceIDs( platform_id, CL_DEVICE_TYPE_DEFAULT, 1, &device_id, &ret_num_devices);

3. 컨텍스트 생성 (선택된 Device를 이용해 Context 생성)

// Create an OpenCL context
cl_context context = clCreateContext( NULL, 1, &device_id, NULL, NULL, &ret);

 

4. 커맨드 큐 생성

// Create a command queue
cl_command_queue command_queue = clCreateCommandQueue(context, device_id, 0, &ret);

5. 메모리 오브젝트 생성

// Create memory buffers on the device for each vector 
cl_mem a_mem_obj = clCreateBuffer(context, CL_MEM_READ_ONLY, LIST_SIZE * sizeof(int), NULL, &ret);
cl_mem b_mem_obj = clCreateBuffer(context, CL_MEM_READ_ONLY, LIST_SIZE * sizeof(int), NULL, &ret);
cl_mem c_mem_obj = clCreateBuffer(context, CL_MEM_WRITE_ONLY, LIST_SIZE * sizeof(int), NULL, &ret);

6. 메모리 버퍼에 복사하기

// Copy the lists A and B to their respective memory buffers
ret = clEnqueueWriteBuffer(command_queue, a_mem_obj, CL_TRUE, 0, LIST_SIZE * sizeof(int), A, 0, NULL, NULL);
ret = clEnqueueWriteBuffer(command_queue, b_mem_obj, CL_TRUE, 0, LIST_SIZE * sizeof(int), B, 0, NULL, NULL);

7. 프로그램 오브젝트 생성

// Create a program from the kernel source
cl_program program = clCreateProgramWithSource(context, 1, (const char **)&source_str, (const size_t *)&source_size, &ret);

8. 커널 컴파일

// Build the program
ret = clBuildProgram(program, 1, &device_id, NULL, NULL, NULL);

9. 커널 오브젝트 생성

// Create the OpenCL kernel
cl_kernel kernel = clCreateKernel(program, "vector_add", &ret);

10. 커널 파라미터 설정

// Set the arguments of the kernel
ret = clSetKernelArg(kernel, 0, sizeof(cl_mem), (void *)&a_mem_obj);
ret = clSetKernelArg(kernel, 1, sizeof(cl_mem), (void *)&b_mem_obj);
ret = clSetKernelArg(kernel, 2, sizeof(cl_mem), (void *)&c_mem_obj);

11. 커널 실행

// Execute the OpenCL kernel on the list
size_t global_item_size = LIST_SIZE; // Process the entire lists
size_t local_item_size = 64; // Divide work items into groups of 64
ret = clEnqueueNDRangeKernel(command_queue, kernel, 1, NULL, &global_item_size, &local_item_size, 0, NULL, NULL);

12. 메모리 오브젝트 로드 (Device로부터 연산이 완료된 Buffer Data 읽기)

// Read the memory buffer C on the device to the local variable C
int *C = (int*)malloc(sizeof(int)*LIST_SIZE);
ret = clEnqueueReadBuffer(command_queue, c_mem_obj, CL_TRUE, 0, LIST_SIZE * sizeof(int), C, 0, NULL, NULL);

13. 오브젝트 해제

// Clean up
ret = clFlush(command_queue);
ret = clFinish(command_queue);
ret = clReleaseKernel(kernel);
ret = clReleaseProgram(program);
ret = clReleaseMemObject(a_mem_obj);
ret = clReleaseMemObject(b_mem_obj);
ret = clReleaseMemObject(c_mem_obj);
ret = clReleaseCommandQueue(command_queue);
ret = clReleaseContext(context);

 

과제 수행중 Trouble Shooting(SGEMM)

OpenCL을 이용해 행렬연산을 하는 코드를 작성하는데 애를 먹었다. host program에서 kernel program으로 전달해주는 메모리 변수를 1차원 배열로 만들어 주어서 그런지 kernel progrma에서도 1차원 배열을 가지고 행렬을 계산했어야했다. 따라서 kernel progrma의 매개변수가 row major 방식의 1차원 배열로 있다고 생각하고 프로그램을 작성하면 되었다.