compilation - como compilar proyecto opencl con kernels

Question

Soy totalmente un principiante en opencl, busqué en Internet y encontré algunas demostraciones de "helloworld" para el proyecto opencl. Por lo general, en este tipo de proyecto mínimo, hay un archivo * .cl que contiene algún tipo de núcleo abierto y un archivo * .c contiene la función principal. Entonces la pregunta es cómo compilar este tipo de proyecto utilizando una línea de comando. Sé que debo usar algún tipo de indicador -lOpenCL en Linux y -framework OpenCL en mac. Pero no tengo idea de vincular el kernel * .cl a mi archivo fuente principal. Gracias por cualquier comentario o enlaces útiles.

Answer 1

Mi respuesta llega cuatro años tarde. Sin embargo, tengo algo para agregar que complementa la respuesta de @ Farzad, como sigue.

Confusamente, en la práctica de OpenCL, el verbo compilar se usa para significar dos cosas diferentes e incompatibles:

• En un uso, compilar significa lo que ya crees que significa. Significa construir en tiempo de compilación, a partir de fuentes * .c para producir objetos * .o para enlaces de tiempo de compilación.
• Sin embargo, en otro uso, y este otro uso puede ser desconocido para usted, compilar los medios para interpretar en tiempo de ejecución, a partir de fuentes * .cl, que producen el código de máquina de la GPU.

Uno sucede en tiempo de construcción. El otro sucede en tiempo de ejecución.

Podría haber sido menos confuso si se hubieran introducido dos verbos diferentes, pero no es así como ha evolucionado la terminología. Convencionalmente, el verbo compilar se usa para ambos.

Si no está seguro, intente este experimento: cambie el nombre de su archivo * .cl para que sus otros archivos de origen no puedan encontrarlo, luego compile.

¿Ver? Se construye bien, ¿no?

Esto se debe a que el archivo * .cl no se consulta en el momento de la compilación. Solo más tarde, cuando intenta ejecutar el ejecutable binario, falla el programa.

Si le ayuda, puede pensar en el archivo * .cl como si fuera un archivo de datos o un archivo de configuración o incluso un script. Tal vez no sea literalmente un archivo de datos, un archivo de configuración o un script, ya que eventualmente se compila en un tipo de código de máquina, pero el código de la máquina es un código de GPU y no se crea a partir del texto del programa * .cl Hasta el tiempo de ejecución. Además, en tiempo de ejecución, su compilador de C como tal no está involucrado. Más bien, es su biblioteca OpenCL la que hace el edificio.

Me tomó bastante tiempo aclarar estos conceptos en mi mente, principalmente porque, como usted, hacía tiempo que estaba familiarizado con las etapas del ciclo de construcción de C / C ++; y, por lo tanto, pensé que sabía lo que significaban las palabras compilar . Una vez que su mente tiene las palabras y los conceptos correctos, la documentación de OpenCL comienza a tener sentido y puede comenzar a trabajar.

Answer 2

En OpenCL, los archivos .cl que contienen códigos de kernel de dispositivos generalmente se compilan y construyen en tiempo de ejecución. Significa que en algún lugar de tu programa OpenCL host, tendrás que compilar y construir el programa de tu dispositivo para poder usarlo. Esta característica permite la máxima portabilidad.

Consideremos un ejemplo que recogí de dos libros. A continuación se muestra un núcleo OpenCL muy simple que agrega dos números de dos arreglos globales y los guarda en otro arreglo global. vector_add_kernel.cl este código en un archivo llamado 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];
}

A continuación se muestra el código de host escrito en C ++ que explota la API de OpenCL C ++. Lo ocl_vector_addition.cpp en un archivo llamado ocl_vector_addition.cpp al lado de donde ocl_vector_addition.cpp mi archivo .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 );
}

Compilo este código en una máquina con Ubuntu 12.04 así:

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

Produce un ocl_vector_addition.o , que cuando ejecuto, muestra un resultado exitoso. Si observa el comando de compilación, verá que no hemos pasado nada sobre nuestro archivo .cl . Solo hemos utilizado el indicador -lOpenCL para habilitar la biblioteca OpenCL para nuestro programa. Además, no se distraiga con el -std=c++11 . Debido a que usé std::unique_ptr en el código del host, tuve que usar esta bandera para una compilación exitosa.

Entonces, ¿dónde se está utilizando este archivo .cl ? Si observa el código de host, encontrará cuatro partes que repito en el siguiente número:

// 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");

En el primer paso, leemos el contenido del archivo que contiene el código de nuestro dispositivo y lo sourceCode en un std::string llamado sourceCode . Luego hacemos un par de la cadena y su longitud y lo guardamos en la source que tiene el tipo cl::Program::Sources . Después de preparar el código, creamos un objeto cl::program llamado program para el context y cargamos el código fuente en el objeto programado. El tercer paso es aquel en el que el código de OpenCL se compila (y vincula) para el device . Dado que el código del dispositivo está integrado en el 3er paso, podemos crear un objeto del kernel llamado vecadd_kernel y asociar el kernel llamado vecadd dentro de nuestro objeto cl::kernel . Esto fue más o menos el conjunto de pasos involucrados en la compilación de un archivo .cl en un programa.

El programa que mostré y expliqué crea el programa del dispositivo desde el código fuente del kernel. Otra opción es utilizar binarios en su lugar. El uso de un programa binario mejora el tiempo de carga de la aplicación y permite la distribución binaria del programa, pero limita la portabilidad ya que los binarios que funcionan bien en un dispositivo pueden no funcionar en otro dispositivo. La creación de programas con código fuente y binarios también se denomina compilación en línea y sin conexión respectivamente (más información here ). Lo omito aquí ya que la respuesta ya es demasiado larga.