2011-09-07 21 views
5

我想只是得到一個基本的程序,使用clCreateProgramWithBinary工作。這是我知道如何使用它而不是一個「真正的」應用程序。如何在OpenCL中使用clCreateProgramWithBinary?

我看到其中一個參數是一個二進制文件列表。我如何去創建一個二進制文件來測試?我有一些測試代碼,它從源代碼創建一個程序,構建並排入它。有沒有在這個過程中的某個時間點創建的二進制文件,我可以輸入到clCreateProgramWithBinary中?

這是我的一些代碼,只是爲了給我一個整體流程的概念。爲簡單起見,我省略了評論和錯誤檢查。

program = clCreateProgramWithSource(clctx, 1, &dumbkernelsource, NULL, &errcode); 
errcode = clBuildProgram(program, env->num_devices, env->device, NULL, NULL, NULL); 
mykernel = clCreateKernel(program, "flops", &errcode); 
errcode = clGetKernelWorkGroupInfo(mykernel, *(env->device), CL_KERNEL_WORK_GROUP_SIZE, sizeof(local), &local, NULL); 
global = num_workgroups * local; 
errcode = clEnqueueNDRangeKernel(commands, mykernel, 1, NULL, &global, &local, 0, NULL, NULL); 

回答

2

編譯程序後,就可以得到它的二進制代碼clGetProgramInfo,然後將其保存到一個文件中。

示例代碼(沒試過編譯,而應是東西沿着這些路線):

program = clCreateProgramWithSource(clctx, 1, &dumbkernelsource, NULL, &errcode); 
errcode = clBuildProgram(program, env->num_devices, env->device, NULL, NULL, NULL); 
int number_of_binaries; 
char **binary; 
int *binary_sizes; 
errcode = clGetProgramInfo(program, CL_PROGRAM_BINARY_SIZES, NULL, 0, &number_of_binaries); 
binary_sizes = new int[number_of_binaries]; 
binary = new char*[number_of_binaries]; 
errcode = clGetProgramInfo(program, CL_PROGRAM_BINARY_SIZES, binary_sizes, number_of_binaries*sizeof(int), &number_of_binaries); 
for (int i = 0; i < number_of_binaries; ++i) binary[i] = new char[binary_sizes[i]]; 
errcode = clGetProgramInfo(program, CL_PROGRAM_BINARIES, binary, number_of_binaries*sizeof(char*), &number_of_binaries); 
0

官方的OpenCL編程指南書有這樣一個很好的例子。還有一個Google代碼項目,opencl-book-sa​​mples,其中包含了該書中的代碼。你正在尋找的例子是here

0

最小可運行例如

編譯從CL C源嵌入向量增量着色器,二進制保存到a.bin,裝載二進制着色器,並運行它:

./a.out 

斷言在完成程序結束。

忽略來自a.bin的CLÇ着色器,負載二進制,並運行它:

./a.out 0 

編譯並執行與:

gcc -ggdb3 -std=c99 -Wall -Wextra a.c -lOpenCL && ./a.out 

測試在Ubuntu 16.10,NVIDIA NVS5400,驅動375.39。

GitHub的上游:https://github.com/cirosantilli/cpp-cheat/blob/b1e9696cb18a12c4a41e0287695a2a6591b04597/opencl/binary_shader.c

#include <assert.h> 
#include <stdio.h> 
#include <stdlib.h> 
#include <string.h> 

#define CL_USE_DEPRECATED_OPENCL_1_2_APIS 
#include <CL/cl.h> 

const char *source = 
    "__kernel void kmain(__global int *out) {\n" 
    " out[get_global_id(0)]++;\n" 
    "}\n" 
; 

#define BIN_PATH "a.bin" 

char* common_read_file(const char *path, long *length_out) { 
    char *buffer; 
    FILE *f; 
    long length; 

    f = fopen(path, "r"); 
    assert(NULL != f); 
    fseek(f, 0, SEEK_END); 
    length = ftell(f); 
    fseek(f, 0, SEEK_SET); 
    buffer = malloc(length); 
    if (fread(buffer, 1, length, f) < (size_t)length) { 
     return NULL; 
    } 
    fclose(f); 
    if (NULL != length_out) { 
     *length_out = length; 
    } 
    return buffer; 
} 

int main(int argc, char **argv) { 
    FILE *f; 
    char *binary; 
    cl_command_queue command_queue; 
    cl_context context; 
    cl_device_id device; 
    cl_int input[] = {1, 2}, errcode_ret, binary_status; 
    cl_kernel kernel, binary_kernel; 
    cl_mem buffer; 
    cl_platform_id platform; 
    cl_program program, binary_program; 
    const size_t global_work_size = sizeof(input)/sizeof(input[0]); 
    int use_cache; 
    long lenght; 
    size_t binary_size; 

    if (argc > 1) { 
     use_cache = !strcmp(argv[1], "0"); 
    } else { 
     use_cache = 0; 
    } 

    /* Get the binary, and create a kernel with it. */ 
    clGetPlatformIDs(1, &platform, NULL); 
    clGetDeviceIDs(platform, CL_DEVICE_TYPE_ALL, 1, &device, NULL); 
    context = clCreateContext(NULL, 1, &device, NULL, NULL, NULL); 
    command_queue = clCreateCommandQueue(context, device, 0, NULL); 
    if (use_cache) { 
     binary = common_read_file(BIN_PATH, &lenght); 
     binary_size = lenght; 
    } else { 
     program = clCreateProgramWithSource(context, 1, &source, NULL, NULL); 
     clBuildProgram(program, 1, &device, "", NULL, NULL); 
     kernel = clCreateKernel(program, "kmain", NULL); 
     clGetProgramInfo(program, CL_PROGRAM_BINARY_SIZES, sizeof(size_t), &binary_size, NULL); 
     binary = malloc(binary_size); 
     clGetProgramInfo(program, CL_PROGRAM_BINARIES, binary_size, &binary, NULL); 
     f = fopen(BIN_PATH, "w"); 
     fwrite(binary, binary_size, 1, f); 
     fclose(f); 
    } 
    binary_program = clCreateProgramWithBinary(
     context, 1, &device, &binary_size, 
     (const unsigned char **)&binary, &binary_status, &errcode_ret 
    ); 
    free(binary); 
    clBuildProgram(binary_program, 1, &device, NULL, NULL, NULL); 
    binary_kernel = clCreateKernel(binary_program, "kmain", &errcode_ret); 

    /* Run the kernel created from the binary. */ 
    buffer = clCreateBuffer(context, CL_MEM_READ_WRITE | CL_MEM_COPY_HOST_PTR, sizeof(input), input, NULL); 
    clSetKernelArg(binary_kernel, 0, sizeof(buffer), &buffer); 
    clEnqueueNDRangeKernel(command_queue, binary_kernel, 1, NULL, &global_work_size, NULL, 0, NULL, NULL); 
    clFlush(command_queue); 
    clFinish(command_queue); 
    clEnqueueReadBuffer(command_queue, buffer, CL_TRUE, 0, sizeof(input), input, 0, NULL, NULL); 

    /* Assertions. */ 
    assert(input[0] == 2); 
    assert(input[1] == 3); 

    /* Cleanup. */ 
    clReleaseMemObject(buffer); 
    clReleaseKernel(kernel); 
    clReleaseKernel(binary_kernel); 
    clReleaseProgram(program); 
    clReleaseProgram(binary_program); 
    clReleaseCommandQueue(command_queue); 
    clReleaseContext(context); 
    return EXIT_SUCCESS; 
} 

我高度推薦cat a.bin,其含有人可讀(和編輯)PTX組件,用於本實施方式。