[开发杂项][opencl]clCreateProgramFromBinary

#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 = (char*)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[2];
    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(2, platform, NULL);
    clGetDeviceIDs(platform[0], CL_DEVICE_TYPE_GPU, 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);
        clGetProgramInfo(program, CL_PROGRAM_BINARY_SIZES, sizeof(size_t), &binary_size, NULL);
        binary = (char*)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);
        kernel = clCreateKernel(program, "kmain", NULL);
    }
    binary_program = clCreateProgramWithBinary(
        context, 1, &device, &binary_size,
        (const unsigned char **)&binary, &binary_status, &errcode_ret
    );
    printf("clCreateProgramWithBinary binary_status=%d, errcode_ret=%d\n", binary_status, errcode_ret);

    free(binary);
    clBuildProgram(binary_program, 1, &device, NULL, NULL, NULL);
    binary_kernel = clCreateKernel(binary_program, "kmain", &errcode_ret);
    printf("clCreateKernel errcode_ret=%d\n", 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);
    printf("ok\n");
    clEnqueueNDRangeKernel(command_queue, binary_kernel, 1, NULL, &global_work_size, NULL, 0, NULL, NULL);
    printf("ok2\n");
    clFlush(command_queue);
    printf("ok3\n");
    clFinish(command_queue);
    printf("ok4\n");
    clEnqueueReadBuffer(command_queue, buffer, CL_TRUE, 0, sizeof(input), input, 0, NULL, NULL);
    printf("ok5\n");

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

    printf("ok6\n");

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