[OpenCL 입문 시리즈 4편] 메모리 관리 기초: 버퍼(Buffer), 이미지(Image), 파라미터(Argument) 다루기

안녕하세요! 지난 글에서는 OpenCL 커널 작성법과 빌드 과정을 살펴보았습니다. 이제 커널에서 다룰 데이터가 어떻게 GPU나 CPU 디바이스 메모리에 올라가는지, 즉 OpenCL 메모리 모델과 그 핵심 요소인 버퍼(Buffer)와 이미지(Image)에 대해 알아볼 차례예요. 또한 커널에 파라미터(인자)를 전달하는 다양한 패턴에 대해서도 간단히 짚어보겠습니다.

이번 글에서는 다음 내용을 다룹니다.

• OpenCL 메모리 모델 개요
• 버퍼(Buffer)와 이미지(Image)의 특징과 사용 예
• 커널 파라미터 설정 방법 심화
• CUDA의 메모리 관리 방식과 비교
• 추가 참고 자료 (유튜브)

1. OpenCL 메모리 모델 개요

OpenCL 프로그램에서는 호스트(Host)가 디바이스(Device)에 연산을 시킵니다. 이 때, 데이터를 디바이스 메모리에 올리려면 보통 다음 단계가 필요하죠.

1. 버퍼나 이미지 객체 생성: clCreateBuffer, clCreateImage 등
2. 호스트에서 디바이스로 데이터 복사: clEnqueueWriteBuffer, clEnqueueWriteImage 등
3. 커널 실행 시 해당 버퍼/이미지를 인자로 설정: clSetKernelArg
4. 커널 실행 후 결과를 디바이스에서 호스트로 복사: clEnqueueReadBuffer, clEnqueueReadImage

이처럼 명시적으로 데이터를 “보내고”, “받는” 과정을 수행해야 합니다. CUDA에서는 cudaMalloc, cudaMemcpy 등을 사용하고, 여기서도 비슷한 느낌이지만, 함수 이름만 다르고 더 일반화되어 있을 뿐이에요.

2. 버퍼(Buffer)

버퍼는 1차원 메모리 객체로, 연속된 메모리 배열을 담는 단순한 구조를 갖습니다. 정수 배열, 실수 배열, 구조체 배열 등의 형태로 사용할 수 있죠. 실제로 지난 글들에서 사용한 예제들도 모두 버퍼를 활용한 것이었습니다.

• 버퍼 생성:여기서 CL_MEM_READ_WRITE, CL_MEM_READ_ONLY, CL_MEM_WRITE_ONLY 등 접근 권한을 지정할 수 있고, CL_MEM_COPY_HOST_PTR를 통해 초기 데이터를 호스트 메모리로부터 복사할 수도 있습니다.
• cl_mem buf = clCreateBuffer(context, CL_MEM_READ_WRITE, sizeof(float)*n, nullptr, &err);
• 데이터 쓰기:host_ptr에서 buf로 데이터를 복사하는 명령을 큐에 넣는 것이고, CL_TRUE로 동기화를 설정하면 이 명령이 끝날 때까지 대기하게 됩니다(비동기 호출도 가능).
• clEnqueueWriteBuffer(queue, buf, CL_TRUE, 0, sizeof(float)*n, host_ptr, 0, nullptr, nullptr);
• 데이터 읽기:
• clEnqueueReadBuffer(queue, buf, CL_TRUE, 0, sizeof(float)*n, host_ptr, 0, nullptr, nullptr);

CUDA와 비교하면 cudaMalloc, cudaMemcpy를 쓰는 것과 개념은 비슷하지만, OpenCL은 플랫폼, 디바이스, 큐를 더 명시적으로 다루기 때문에 코드가 조금 길어질 수 있어요.

3. 이미지(Image)

이미지는 2차원 또는 3차원 텍스처 같은 형태의 메모리 객체입니다. GPU에서 이미지 단위를 다룰 때는 텍스처 캐시 최적화를 받을 수 있어 이미지 객체를 쓰는 것이 유리할 때도 있어요. 특히 이미지 처리 애플리케이션이라면, 픽셀 단위 접근과 필터링에 이미지 객체를 활용하는 것이 일반적입니다.

이미지는 다음과 같이 생성합니다.

cl_image_format format;
format.image_channel_order = CL_RGBA;
format.image_channel_data_type = CL_UNORM_INT8;

cl_image_desc desc = {0};
desc.image_type = CL_MEM_OBJECT_IMAGE2D;
desc.image_width = width;
desc.image_height = height;

cl_mem img = clCreateImage(context, CL_MEM_READ_WRITE,
                           &format, &desc, nullptr, &err);

이미지에 데이터를 쓰거나 읽을 때는 clEnqueueWriteImage, clEnqueueReadImage를 사용합니다. 커널 코드 안에서는 read_imagef, write_imagef 같은 전용 함수로 이미지를 다루게 됩니다. CUDA에도 텍스처와 표면 메모리 개념이 있는데, OpenCL 이미지 객체는 그와 유사하지만 좀 더 표준화된 형태로 제공하는 셈이에요.

4. 커널 파라미터 설정

지난 글에서도 clSetKernelArg를 이용해 커널 인자를 설정하는 과정을 살펴봤습니다. 버퍼나 이미지 객체를 커널 인자로 줄 때는 cl_mem 타입의 포인터를 넘기게 되고, 일반적인 스칼라 타입(예: int, float)은 값 자체를 인자로 직접 설정합니다.

clSetKernelArg(kernel, 0, sizeof(cl_mem), &bufA);
clSetKernelArg(kernel, 1, sizeof(cl_mem), &bufB);
clSetKernelArg(kernel, 2, sizeof(int), &n);

커널 코드에서는 __kernel void my_kernel(__global float* A, __global float* B, int n)처럼 인자를 선언한 순서대로 인덱스를 맞춰 넣으면 됩니다.

여기서 __global 키워드는 버퍼가 GPU 메모리에 있음을 나타내고, 이미지 객체는 __read_only image2d_t 와 같이 선언하게 됩니다. 인자 전달은 CUDA에서 함수를 호출하듯이 파라미터를 넘기는 것과 유사하지만, explicit하게 clSetKernelArg를 통해 하나하나 설정해야 한다는 점이 다릅니다.

5. CUDA와 비교하기

• CUDA:
  • cudaMalloc, cudaMemcpy로 호스트-디바이스 간 데이터 전송
  • 텍스처/서피스 메모리도 있지만, 주로 전용 API나 텍스처 참조를 통한 접근
  • 커널 호출 시 직접 파라미터를 넘겨줄 수도 있고, 런칭 시 <<< >>> 구문 안에서 설정
• OpenCL:
  • clCreateBuffer, clEnqueueWriteBuffer, clEnqueueReadBuffer 등 조금 더 장황한 API
  • 이미지 메모리 객체를 통한 2D/3D 데이터 최적화 가능
  • clSetKernelArg로 모든 커널 인자를 명시적으로 설정

결국 둘 다 비슷한 개념을 다른 형태로 제공하는 것이며, OpenCL은 다양한 디바이스 지원을 위해 더 범용적이고 유연한 방식을 취하는 반면, CUDA는 NVIDIA GPU에 특화되어 있어 상대적으로 단순하고 깔끔합니다.

6. 마무리

이번 글에서는 OpenCL에서 데이터 전송 및 메모리 관리의 기초를 다뤄봤습니다. 버퍼와 이미지 개념을 이해하면, 어떠한 데이터가 디바이스 메모리로 넘어가고 돌아오는지 명확히 파악할 수 있고, 커널 인자 설정 과정도 자연스럽게 이해할 수 있죠. 다음 글에서는 이렇게 준비한 메모리를 실제로 커널과 어떻게 연계하는지, 그리고 C++20/23 스타일 래퍼를 이용해 코드를 깔끔하게 정리하는 방법을 살펴볼 예정입니다.

유용한 링크 & 리소스

• OpenCL 공식 홈페이지 : 메모리 객체 관련 표준 문서
• Khronos Group GitHub : OpenCL 관련 예제 코드
• C++ Reference : 최신 C++ 문법 참고용