Запуск ядра OpenCL на нескольких графических процессорах?
Прямо сейчас я запрограммировал создание нескольких алгоритмов, работающих параллельно на одном графическом процессоре, но у всех них возникает одна и та же проблема, когда я пытаюсь выполнить их на нескольких графических процессорах (например, 3). Проблема заключается в том, что код, выполняемый на одном графическом процессоре, выполняется точно так же на 3 графических процессорах (не быстрее). Я пытался выполнить с большим количеством данных, пытался выполнить разные задачи, ничего не помогло. Наконец, я попытался выполнить простейшую задачу, такую как sum sum, и все еще получил эту ужасную ошибку. Вот почему я не верю, что это проблема конкретного алгоритма, и я чувствую, что в моем коде есть ошибка (или даже в моем подходе к распараллеливанию кода на нескольких графических процессорах).
Вот заголовочный файл для моего класса Parallel.cpp:
#ifndef PARALLEL_H
#define PARALLEL_H
#define __NO_STD_VECTOR // Use cl::vector and cl::string and
#define __NO_STD_STRING // not STL versions, more on this later
#include <CL/cl.h>
class Parallel
{
public:
Parallel();
int executeAttachVectorsKernel(int*, int*, int*, int);
static void getMaxWorkGroupSize(int*, int*, int*);
virtual ~Parallel();
protected:
private:
char* file_contents(const char*, int*);
void getShortInfo(cl_device_id);
int init(void);
cl_platform_id platform;
cl_device_id* devices;
cl_uint num_devices;
cl_command_queue* queues;
int* WGSizes;
int* WGNumbers;
cl_context context;
cl_program program;
cl_kernel kernel;
cl_mem input1;
cl_mem input2;
cl_mem output;
};
#endif // PARALLEL_H
Вот метод инициализации init:
int Parallel::init() {
cl_int err;
//Connect to the first platfrom
err = clGetPlatformIDs(1, &platform, NULL);
if (err != CL_SUCCESS) {
cerr << "Error occured while executing clGetPlatformIDs" << endl;
return EXIT_FAILURE;
}
//Get devices number
err = clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, 0, NULL, &num_devices);
if (err != CL_SUCCESS) {
cerr << "Error: Failed to create a device group:" << endl;
return EXIT_FAILURE;
}
cout << "NUM DEVICES =" << num_devices << endl;
devices = new cl_device_id[num_devices];
//Get all the GPU devices
err = clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, num_devices, devices, NULL);
//Create one context for all the devices
context = clCreateContext(NULL, num_devices, devices, NULL, NULL, &err);
if (!context) {
cerr << "Error: Failed to create a compute context!" << endl;
return EXIT_FAILURE;
}
queues = new cl_command_queue[num_devices];
WGNumbers = new int[num_devices];
WGSizes = new int[num_devices];
for(int i = 0; i < num_devices; i++) {
//Create a command queue for every device
queues[i] = clCreateCommandQueue(context, devices[i], 0, &err);
if (!queues[i]) {
cerr << "Error: Failed to create a command commands!" << endl;
return EXIT_FAILURE;
}
cl_ulong temp;
clGetDeviceInfo(devices[i], CL_DEVICE_MAX_WORK_GROUP_SIZE, sizeof(temp), &temp, NULL);
WGSizes[i] = (int)temp;
clGetDeviceInfo(devices[i], CL_DEVICE_MAX_WORK_ITEM_SIZES, sizeof(temp), &temp, NULL);
WGNumbers[i] = (int)temp;
}
//Translate kernel code into chars
int pl;
size_t program_length;
string path = "./kernel/kernel_av.cl";
char* cSourceCL = file_contents(path.c_str(), &pl);
program_length = (size_t)pl;
//Create a program
program = clCreateProgramWithSource(context, 1,
(const char **) &cSourceCL, &program_length, &err);
if (!program) {
cerr << "Error: Failed to create compute program!" << endl;
return EXIT_FAILURE;
}
//Create an executable
err = clBuildProgram(program, 0, NULL, NULL, NULL, NULL);
if (err != CL_SUCCESS)
{
size_t len;
char buffer[2048];
cerr << "Error: Failed to build program executable!" << endl;
exit(1);
}
// Create the compute kernel in the program
kernel = clCreateKernel(program, "calculate2dim", &err);
if (err != CL_SUCCESS)
{
cerr << "Error: Failed to create compute kernel!" << endl;
exit(1);
}
}
Метод, который выполняет ядро, находится здесь:
int Parallel::executeAttachVectorsKernel(int* data1, int* data2, int* results, int vectors_num) {
cl_int err;
size_t global; // global domain size for our calculation
size_t local; // local domain size for our calculation
int partition = vectors_num/num_devices;
unsigned int count = partition;
input1 = clCreateBuffer(context, CL_MEM_READ_ONLY, sizeof(int) * count, NULL, NULL);
input2 = clCreateBuffer(context, CL_MEM_READ_ONLY, sizeof(int) * count, NULL, NULL);
output = clCreateBuffer(context, CL_MEM_WRITE_ONLY, sizeof(int) * count, NULL, NULL);
if (!input1 || !input2 || !output) {
cerr << "Error: Failed to allocate device memory!" << endl;
exit(1);
}
int** data1_apart = new int*[num_devices];
int** data2_apart = new int*[num_devices];
int** results_apart = new int*[num_devices];
for(int i = 0; i < num_devices; i++) {
cout << "Executing parallel part on GPU " << i + 1 << endl;
cout << "Partition size = " << partition << endl;
data1_apart[i] = new int[partition];
data2_apart[i] = new int[partition];
results_apart[i] = new int[partition];
for(int j = i*partition, k = 0; k < partition; j++, k++) {
data1_apart[i][k] = data1[j];
data2_apart[i][k] = data2[j];
}
//Transfer the input vector into device memory
err = clEnqueueWriteBuffer(queues[i], input1,
CL_TRUE, 0, sizeof(int) * count,
data1_apart[i], 0, NULL, NULL);
err = clEnqueueWriteBuffer(queues[i], input2,
CL_TRUE, 0, sizeof(int) * count,
data2_apart[i], 0, NULL, NULL);
if (err != CL_SUCCESS)
{
cerr << "Error: Failed to write to source array!" << endl;
exit(1);
}
int parameter4 = count/WGNumbers[i];
//Set the arguments to the compute kernel
err = 0;
err = clSetKernelArg(kernel, 0, sizeof(cl_mem), &input1);
err |= clSetKernelArg(kernel, 1, sizeof(cl_mem), &input2);
err |= clSetKernelArg(kernel, 2, sizeof(cl_mem), &output);
err |= clSetKernelArg(kernel, 3, sizeof(int), ¶meter4);
if (err != CL_SUCCESS)
{
cerr << "Error: Failed to set kernel arguments! " << err << endl;
exit(1);
}
global = WGNumbers[i];
local = WGSizes[i];
if(local > global) {
local = global;
}
cout << "global = " << global << " local = " << local << endl;
err = clEnqueueNDRangeKernel(queues[i], kernel,
1, NULL, &global, &local,
0, NULL, NULL);
if (err)
{
cerr << "Error: Failed to execute kernel!" << endl;
return EXIT_FAILURE;
}
}
for(int i = 0; i < num_devices; i++) {
//Wait for all commands to complete
clFinish(queues[i]);
//Read back the results from the device to verify the output
err = clEnqueueReadBuffer(queues[i], output,
CL_TRUE, 0, sizeof(int) * count,
results_apart[i], 0, NULL, NULL );
if (err != CL_SUCCESS)
{
cerr << "Error: Failed to read output array! " << err << endl;
exit(1);
}
for(int j = 0; j < partition; j++) {
results[i*partition + j] = results_apart[i][j];
}
delete [] data1_apart[i];
delete [] data2_apart[i];
delete [] results_apart[i];
}
clReleaseMemObject(input1);
clReleaseMemObject(input2);
clReleaseMemObject(output);
delete [] data1_apart;
delete [] data2_apart;
}
До публикации этого вопроса в stackru я боролся с этой проблемой в течение 2-3 недель, и теперь мне действительно нужна чья-то помощь, поэтому я буду очень признателен за любые мысли и ответы!
3 ответа
Вот что я думаю, что происходит. Вы вызываете clEnqueueNDRangeKernel один раз для каждого участвующего устройства opencl. На данный момент ни одно из ядер не запустилось, потому что clFlush не был вызван. Далее вы делаете clFinish для каждой очереди. При первом вызове clFinish запускается первая рабочая группа из очереди. Это также ждет, пока это закончится. По завершении первой рабочей группы clFinish возвращает управление вашему приложению. Затем ваше приложение вызывает clFinish для следующей очереди. Это запускает второй рабочий раствор, а также ожидает его завершения. Так что работа идет последовательно. Решение может быть таким же простым, как вызов clFush сразу после каждого вызова clEnqueueNDRangeKernel. Вот как ведет себя моя система AMD. Я приведу рабочий пример в ближайшее время.
Все ваши устройства работают с одинаковыми буферами. Данные будут перемещаться между устройствами при запуске ядер. Без правильной синхронизации результаты будут неопределенными.
Если возможно, рассмотрите возможность выделения отдельного набора буферов для каждого устройства.
Какие графические процессоры вы используете? У меня есть GTX590, который появляется на двух устройствах с графическим процессором. Когда я пытался запустить его на обоих устройствах, казалось, что оно дожидается завершения каждого устройства, прежде чем оно перейдет к следующему (хотя это и не предполагалось). Я не знаю, исправил ли это Nvidia.
Читая некоторые сообщения, я думаю, что на сайте Nvidia в то время я читал кое-что о Nvidia, предлагая создать отдельные контексты для каждого устройства и запускать их в разных потоках. Это то, что я сделал, и это прекрасно работает. Для этого я использовал pthreads (или SDL_threads). Это довольно легко настроить.