OpenCL 기초

유호준·2022년 5월 12일

목록 보기

2/3

OpenCL의 예제들은 대부분이 C 스타일의 코드들이다. 나는 C++ Wrapper API로 C++ 스타일로 작성을 해보려고한다.

Platform

첫번째로, OpenCL을 실행할 플랫폼을(Ex: Intel, Nvidia, etc...) 찾아야한다.

cl_int errNum;
std::vector<Platform> platforms;
errNum = Platform::get(&platforms);

if (errNum != CL_SUCCESS || platforms.size() <= 0) {
	cerr << "Failed to find any opencl platforms" << endl;
	return;
}

cout << "Number of platforms " << platforms.size() << endl;
    
for (Platform platform : platforms) {
	string platform_profile, platform_version, platform_vendor;
	errNum = platform.getInfo(CL_PLATFORM_PROFILE, &platform_profile);
	errNum |= platform.getInfo(CL_PLATFORM_VERSION, &platform_version);
	errNum |= platform.getInfo(CL_PLATFORM_VENDOR, &platform_vendor);

	if (errNum != CL_SUCCESS) {
		cerr << "Failed to get info about platform" << endl;
		return;
	}

	cout << platform_profile << ":" << platform_version << ":" << platform_vendor << endl;
}

Device

다음은 원하는 플랫폼의 어떤 기기를 사용할 건지 찾는다. 여기서는 그냥 첫번째 GPU기기를 사용할 것이다.

for (Platform platform : platforms) {
	std::vector<Device> devices;
	err_num = platform.getDevices(CL_DEVICE_TYPE_GPU, &devices);
	if (err_num != CL_SUCCESS && err_num == CL_DEVICE_NOT_FOUND) {
		checkErr(err_num, "platform.getDevices");
	}
	else if (devices.size() > 0) {
		device = devices[0];
		break;
	}
}

코드는 간단하다. 플랫폼을 반복문으로 확인하면서 GPU 기기를 찾았다면 반복문을 탈출하는 코드이다.

checkErr는 에러를 확인하는 함수이다.

Context

Context는 OpenCL의 문맥이다. 커널이 실행되는 곳을 의미한다.

context = Context(device, NULL,NULL, &err_num);

CommandQueue

CommandQueue는 OpenCL의 Queue이다.

CommandQueue queue = CommandQueue(context, device, 0, &err_num);

Program

Program은 OpenCL의 커널들을 모아놓은 객체이다. 소스코드를 문자열로 읽어서 넣어주면 된다.

ifstream kerenlFile(file_name, ios::in);
if (!kerenlFile.is_open()) {
	throw "Failed to open file for reading " + file_name;
}
ostringstream oss;
oss << kerenlFile.rdbuf();
string srcStdStr = oss.str();
program = Program(context, srcStdStr.c_str(),false,&errNum);

program.build({ device });
if (errNum != CL_SUCCESS) {
	throw "Failed to build Program " +  program.getBuildInfo<CL_PROGRAM_BUILD_LOG>(device);
}

Kernel

Kernel은 우리가 병렬처리를 할 함수라고 생각하면 된다.

kernel = Kernel(program, "hello", &errNum);

여기서 Program의 생성에 사용된 코드는 아래와 같다. 단순한 배열의 덧셈을 수행하는 코드이다.

void kernel hello(global const float *a, global const float *b, global float* result){
	int gid = get_global_id(0);
	result[gid] = a[gid] + b[gid];
}

위 함수는 병렬처리를 하게 될 것이므로 자기가 처리할 것이 어떤 것인지를 찾아야 한다. 반복문의 index 느낌으로 생각하면 편하다. 이제 준비는 끝났고 실행을 하면 된다.

global은 OpenCL내부의 주소 공간을 의미한다.

Buffer

hello kernel을 보면 a, b, result 세가지의 매개변수가 들어간다. 실행을 하기 전 이 공간을 확보를 해줘야한다. OpenCL에서는 Buffer객체로 확보를 하게된다.

memObjects[0] = Buffer::Buffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, 10 * sizeof(float), a, &errNums[0]);
memObjects[1] = Buffer::Buffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, 10 * sizeof(float), b, &errNums[1]);
memObjects[2] = Buffer::Buffer(context, CL_MEM_READ_WRITE, 10 * sizeof(float), NULL, &errNums[2]);

CL_MEM_READ_ONLY는 읽기 전용, CL_MEM_READ_WRITE는 읽기 쓰기

실행

이제 진짜 실행이다!

errNum = kernel.setArg(0, memObjects[0]);
errNum |= kernel.setArg(1, memObjects[1]);
errNum |= kernel.setArg(2, memObjects[2]);

우리가 확보한 Buffer를 Kernel에 매개변수로 넣어준다. 0, 1, 2는 매개변수의 순서를 의미한다.

errNum = queue.enqueueNDRangeKernel(kernel, NullRange, NDRange(10));
errNum = queue.enqueueReadBuffer(memObjects[2], CL_TRUE, 0, sizeof(float) * 10, result);

실행은 CommandQueue를 통해 실행한다. NDRange(10)은 global work-items의 크기를 의미를 하는데 여기서는 길이 10의 배열을 사용하기 때문에 10을 넣는다.

실행 결과 역시 CommandQueue를 통해 읽어온다.

결과

for (int i = 0; i < 10; i++) {
	a[i] = i;
	b[i] = i * 2;
}

전체 코드

#include<CL/cl2.hpp>
#include<iostream>
#include<vector>
#include<fstream>
#include<sstream>

using namespace std;
using namespace cl;

int main(void) {
	Context context;
	CommandQueue queue;
	Program program;
	Device device;
	Kernel kernel;
	cl_int errNum;
	Buffer memObjects[3];

	//초기 설정
	try {
		context = create_context();
		queue = create_command_queue(context, device);
		program = create_program(context, device, "hello.cl");
	}
	catch (string errMsg) {
		cerr << errMsg << endl;
		exit(EXIT_FAILURE);
	}

	kernel = Kernel(program, "hello", &errNum);
	if (errNum != CL_SUCCESS) {
		cerr << "Failed to create kernel" << endl;
		exit(EXIT_FAILURE);
	}

	float result[10];
	float a[10];
	float b[10];

	for (int i = 0; i < 10; i++) {
		a[i] = i;
		b[i] = i * 2;
	}

	if (!create_buffer_objects(context, memObjects, a, b)) {
		exit(EXIT_FAILURE);
	}

	errNum = kernel.setArg(0, memObjects[0]);
	errNum |= kernel.setArg(1, memObjects[1]);
	errNum |= kernel.setArg(2, memObjects[2]);

	if (errNum != CL_SUCCESS) {
		cerr << "Error setting kerenl args" << endl;
		exit(EXIT_FAILURE);
	}

	errNum = queue.enqueueNDRangeKernel(kernel, NullRange, NDRange(10));
	if (errNum != CL_SUCCESS) {
		cerr << "Error queuing kernel for execution" << endl;
		exit(EXIT_FAILURE);
	}

	errNum = queue.enqueueReadBuffer(memObjects[2], CL_TRUE, 0, sizeof(float) * 10, result);
	if (errNum != CL_SUCCESS) {
		cerr << "Error reading result buffer" << endl;
		exit(EXIT_FAILURE);
	}

	for (int i = 0; i < 10; i++) {
		cout << result[i] << " ";
	}
	cout << endl;

	exit(EXIT_SUCCESS);
}

Context create_context() noexcept(false){//첫 번째로 사용 가능한 플랫폼에서 생성
	cl_int errNum;
	Context context;
	std::vector<Platform> platforms;
	errNum = Platform::get(&platforms);

	if (errNum != CL_SUCCESS || platforms.size() == 0) {
		throw "Failed to find any OpenCL Platforms";
	}

	context = Context(CL_DEVICE_TYPE_GPU,NULL,NULL,&errNum);
	if (errNum != CL_SUCCESS) {
		throw "Could not create GPU context";
		
	}
	return context;
}

CommandQueue create_command_queue(Context context, Device &device) noexcept(false){
	cl_int errNum;
	CommandQueue queue;
	std::vector<Device> devices;
	string device_name, device_hardware_version, device_software_version, device_opencl_c_version;

	errNum = context.getInfo(CL_CONTEXT_DEVICES, &devices);
	if (errNum != CL_SUCCESS) {
		throw "Failed call to context getInfo";
	}
	if (devices.size() <= 0) {
		throw "No device available";
	}
	
	queue = CommandQueue::CommandQueue(context, devices[0],0,&errNum);
	if (errNum != CL_SUCCESS) {
		throw "Failed to create commandQueue for device 0";
	}
	device = devices[0];
	device.getInfo(CL_DEVICE_NAME, &device_name);
	device.getInfo(CL_DEVICE_VERSION, &device_hardware_version);
	device.getInfo(CL_DRIVER_VERSION, &device_software_version);
	device.getInfo(CL_DEVICE_OPENCL_C_VERSION, &device_opencl_c_version);
	cout << device_name << ":" << device_hardware_version << ":" << device_software_version << ":" << device_opencl_c_version << endl;

	return queue;
}

Program create_program(Context context, Device device, const string file_name) noexcept(false){
	cl_int errNum;
	Program program;
	
	ifstream kerenlFile(file_name, ios::in);
	if (!kerenlFile.is_open()) {
		throw "Failed to open file for reading " + file_name;
	}
	ostringstream oss;
	oss << kerenlFile.rdbuf();
	string srcStdStr = oss.str();
	program = Program(context, srcStdStr.c_str(),false,&errNum);

	if (errNum != CL_SUCCESS) {
		throw "Failed to create program";
	}

	program.build({ device });
	if (errNum != CL_SUCCESS) {
		throw "Failed to build Program " +  program.getBuildInfo<CL_PROGRAM_BUILD_LOG>(device);
	}

	return program;
}

bool create_buffer_objects(Context context, Buffer memObjects[3], float* a, float* b) {
	cl_int errNums[3];
	memObjects[0] = Buffer::Buffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, 10 * sizeof(float), a, &errNums[0]);
	memObjects[1] = Buffer::Buffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, 10 * sizeof(float), b, &errNums[1]);
	memObjects[2] = Buffer::Buffer(context, CL_MEM_READ_WRITE, 10 * sizeof(float), NULL, &errNums[2]);

	if (errNums[0] != CL_SUCCESS || errNums[1] != CL_SUCCESS || errNums[2] != CL_SUCCESS) {
		cerr << "Error creating memory objects" << endl;
		return false;
	}
	return true;
}