Как скомпилировать opencl-проект с ядрами

Ответ 1

В OpenCL файлы .cl, содержащие коды ядра устройства, обычно компилируются и создаются во время выполнения. Это означает, что где-то в вашей программе OpenCL хоста вам придется скомпилировать и создать свою программу устройств, чтобы иметь возможность ее использовать. Эта функция обеспечивает максимальную переносимость.

Рассмотрим пример, который я собрал из двух книг. Ниже представлено очень простое ядро ​​OpenCL, добавляющее два числа из двух глобальных массивов и сохранение их в другом глобальном массиве. Я сохраняю этот код в файле с именем vector_add_kernel.cl.

kernel void vecadd( global int* A, global int* B, global int* C ) {
    const int idx = get_global_id(0);
    C[idx] = A[idx] + B[idx];
}

Ниже приведен код узла, написанный на С++, который использует OpenCL С++ API. Я сохраняю его в файле с именем ocl_vector_addition.cpp рядом с тем, где я сохранил файл .cl.

#include <iostream>
#include <fstream>
#include <string>
#include <memory>
#include <stdlib.h>

#define __CL_ENABLE_EXCEPTIONS
#if defined(__APPLE__) || defined(__MACOSX)
#include <OpenCL/cl.cpp>
#else
#include <CL/cl.hpp>
#endif

int main( int argc, char** argv ) {

    const int N_ELEMENTS=1024*1024;
    unsigned int platform_id=0, device_id=0;

    try{
        std::unique_ptr<int[]> A(new int[N_ELEMENTS]); // Or you can use simple dynamic arrays like: int* A = new int[N_ELEMENTS];
        std::unique_ptr<int[]> B(new int[N_ELEMENTS]);
        std::unique_ptr<int[]> C(new int[N_ELEMENTS]);

        for( int i = 0; i < N_ELEMENTS; ++i ) {
            A[i] = i;
            B[i] = i;
        }

        // Query for platforms
        std::vector<cl::Platform> platforms;
        cl::Platform::get(&platforms);

        // Get a list of devices on this platform
        std::vector<cl::Device> devices;
        platforms[platform_id].getDevices(CL_DEVICE_TYPE_GPU|CL_DEVICE_TYPE_CPU, &devices); // Select the platform.

        // Create a context
        cl::Context context(devices);

        // Create a command queue
        cl::CommandQueue queue = cl::CommandQueue( context, devices[device_id] );   // Select the device.

        // Create the memory buffers
        cl::Buffer bufferA=cl::Buffer(context, CL_MEM_READ_ONLY, N_ELEMENTS * sizeof(int));
        cl::Buffer bufferB=cl::Buffer(context, CL_MEM_READ_ONLY, N_ELEMENTS * sizeof(int));
        cl::Buffer bufferC=cl::Buffer(context, CL_MEM_WRITE_ONLY, N_ELEMENTS * sizeof(int));

        // Copy the input data to the input buffers using the command queue.
        queue.enqueueWriteBuffer( bufferA, CL_FALSE, 0, N_ELEMENTS * sizeof(int), A.get() );
        queue.enqueueWriteBuffer( bufferB, CL_FALSE, 0, N_ELEMENTS * sizeof(int), B.get() );

        // Read the program source
        std::ifstream sourceFile("vector_add_kernel.cl");
        std::string sourceCode( std::istreambuf_iterator<char>(sourceFile), (std::istreambuf_iterator<char>()));
        cl::Program::Sources source(1, std::make_pair(sourceCode.c_str(), sourceCode.length()));

        // Make program from the source code
        cl::Program program=cl::Program(context, source);

        // Build the program for the devices
        program.build(devices);

        // Make kernel
        cl::Kernel vecadd_kernel(program, "vecadd");

        // Set the kernel arguments
        vecadd_kernel.setArg( 0, bufferA );
        vecadd_kernel.setArg( 1, bufferB );
        vecadd_kernel.setArg( 2, bufferC );

        // Execute the kernel
        cl::NDRange global( N_ELEMENTS );
        cl::NDRange local( 256 );
        queue.enqueueNDRangeKernel( vecadd_kernel, cl::NullRange, global, local );

        // Copy the output data back to the host
        queue.enqueueReadBuffer( bufferC, CL_TRUE, 0, N_ELEMENTS * sizeof(int), C.get() );

        // Verify the result
        bool result=true;
        for (int i=0; i<N_ELEMENTS; i ++)
            if (C[i] !=A[i]+B[i]) {
                result=false;
                break;
            }
        if (result)
            std::cout<< "Success!\n";
        else
            std::cout<< "Failed!\n";

    }
    catch(cl::Error err) {
        std::cout << "Error: " << err.what() << "(" << err.err() << ")" << std::endl;
        return( EXIT_FAILURE );
    }

    std::cout << "Done.\n";
    return( EXIT_SUCCESS );
}

Я компилирую этот код на машине с Ubuntu 12.04 следующим образом:

g++ ocl_vector_addition.cpp -lOpenCL -std=c++11 -o ocl_vector_addition.o

Он создает ocl_vector_addition.o, который при запуске показывает успешный вывод. Если вы посмотрите на команду компиляции, вы увидите, что мы ничего не передали о нашем файле .cl. Мы использовали флаг -lOpenCL, чтобы включить библиотеку OpenCL для нашей программы. Кроме того, не отвлекайтесь на команду -std=c++11. Поскольку я использовал std::unique_ptr в главном коде, мне пришлось использовать этот флаг для успешного компиляции.

Итак, где используется этот .cl файл? Если вы посмотрите на код хоста, вы найдете четыре части, которые я повторяю ниже:

        //1. Read the program source
        std::ifstream sourceFile("vector_add_kernel.cl");
        std::string sourceCode( std::istreambuf_iterator<char>(sourceFile), (std::istreambuf_iterator<char>()));
        cl::Program::Sources source(1, std::make_pair(sourceCode.c_str(), sourceCode.length()));

        //2. Make program from the source code
        cl::Program program=cl::Program(context, source);

        //3. Build the program for the devices
        program.build(devices);

        //4. Make kernel
        cl::Kernel vecadd_kernel(program, "vecadd");

На первом этапе мы читаем содержимое файла, содержащего наш код устройства, и помещаем его в std::string с именем sourceCode. Затем мы создаем пару строк и их длину и сохраняем ее до source, которая имеет тип cl::Program::Sources. После того, как мы подготовили код, мы создаем объект cl::program с именем program для context и загружаем исходный код в объект программы. Третий шаг - это тот, в котором код OpenCL компилируется (и привязывается) к device. Поскольку код устройства построен на 3-м шаге, мы можем создать объект ядра с именем vecadd_kernel и связать ядро ​​с именем vecadd внутри него с нашим объектом cl::kernel. Это был довольно много шагов, связанных с компиляцией файла .cl в программе.

Программа, которую я показал и объяснил, создает программу устройства из исходного кода ядра. Другой вариант - вместо этого использовать двоичные файлы. Использование двоичной программы увеличивает время загрузки приложений и позволяет бинарное распределение программы, но ограничивает переносимость, поскольку двоичные файлы, которые работают нормально на одном устройстве, могут не работать на другом устройстве. Создание программы с использованием исходного кода и двоичного файла также называется автономной и онлайн-компиляцией (подробнее здесь). Я пропущу его здесь, так как ответ уже слишком длинный.