3

I have the following very simple kernel in OpenCL

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

I created a C++ program to load the kernel from a binary created from its source. The binary loads correctly (CL_SUCCESS), but does not display the correct result for the input. It displays changing garbage values like so

result: 538976310 538976288 538976288 538976288 538976288 790634528 796160111 1702129257 1886334828 1818455653

inline cl::Program CreateProgramFromBinary(cl::Context context,const std::vector<cl::Device> devices, const char* fileName)
{
    std::ifstream file(fileName,  std::ios::binary | std::ios::in | std::ios::ate);

    uint32_t size = file.tellg();
    file.seekg(0, std::ios::beg);
    char* buffer = new char[size];
    file.read(buffer, size);
    file.close();
    cl::Program::Binaries bin{{buffer, size}};

    std::vector<cl_int> binaryStatus;
    cl_int err = 0;
    cl::Program program = cl::Program{context, devices, bin, &binaryStatus, &err};

    if(err != CL_SUCCESS) {
       std::cout<<" Error loading"<< err<<  "\n";
        exit(1);
    }
    for (std::vector<cl_int>::const_iterator bE = binaryStatus.begin(); bE != binaryStatus.end(); bE++) {
        std::cout<< *bE <<std::endl;
    }
    std::cout<<"No Error loading"<< err<<  "\n";
    delete[] buffer;
    return program;
}

int main(int argc, char** argv)
{
    std::vector<cl::Device> devices= loadDevices();
    cl::Context context{devices};

    std::cout << "Save program binary for future run..." << std::endl;
    //cl::Program program = CreateBinaryFromProgram(context, devices, "HelloWorld.cl", "HelloWorld.cl.bin");
    //CreateBinaryFromProgram(context, devices, "HelloWorld.cl", "HelloWorld.cl.bin");


    std::cout << "Reading from binary..." << std::endl;
    cl::Program program = CreateProgramFromBinary(context, devices, "HelloWorld.cl.bin");

    std::cout << "Running Program..." << std::endl;
    cl::Buffer buffer_A(context,CL_MEM_READ_WRITE,sizeof(int)*10);
    cl::Buffer buffer_B(context,CL_MEM_READ_WRITE,sizeof(int)*10);
    cl::Buffer buffer_C(context,CL_MEM_READ_WRITE,sizeof(int)*10);

    int A[] = {0, 1, 2, 3, 4, 5, 6, 7, 8, 9};
    int B[] = {0, 1, 2, 3, 4, 5, 6, 7, 8, 9};

    //create queue to which we will push commands for the device.
    cl::CommandQueue queue(context,devices[0]);

    //write arrays A and B to the device
    queue.enqueueWriteBuffer(buffer_A,CL_TRUE,0,sizeof(int)*10,A);
    queue.enqueueWriteBuffer(buffer_B,CL_TRUE,0,sizeof(int)*10,B);


    //run the kernel
    cl::Kernel kernel_add=cl::Kernel(program,"simple_add");
    kernel_add.setArg(0,buffer_A);
    kernel_add.setArg(1,buffer_B);
    kernel_add.setArg(2,buffer_C);
    queue.enqueueNDRangeKernel(kernel_add,cl::NullRange,cl::NDRange(10),cl::NullRange);
    queue.finish();

    int C[10];
    //read result C from the device to array C
    queue.enqueueReadBuffer(buffer_C,CL_TRUE,0,sizeof(int)*10,C);

    std::cout<<" result: \n";
    for(int i=0;i<10;i++)
        std::cout<<C[i]<<" ";
    std::cout << "\n";
    return 0;
}

Loading this program directly from the CL file however, results in the correct output of the program. Is the binary I've loaded somehow different from the CL file?


EDIT:

How I created the Binary

inline cl::Program CreateBinaryFromProgram(const cl::Context context,const std::vector<cl::Device> devices, const char* readFileName, const char* writeFileName)
{
    std::ifstream file(readFileName, std::ios::binary| std::ios::ate | std::ios::in);

    uint32_t size = file.tellg();
    file.seekg(0, std::ios::beg);
    char* buffer = new char[size];

    file.read(buffer, size);
    file.close();

    cl::Program::Sources sources;

    // kernel calculates for each element C=A+B
    std::string kernel_code(buffer);
    sources.push_back({kernel_code.c_str(),kernel_code.length()});
    cl::Program program{context,sources};
    if(program.build(devices)!=CL_SUCCESS){
       std::cout<<" Error building: "<<program.getBuildInfo<CL_PROGRAM_BUILD_LOG>(devices[0])<<"\n";
        exit(1);
    }
    std::vector<size_t> output_sizes = program.getInfo<CL_PROGRAM_BINARY_SIZES>();
    std::vector<char*> output = program.getInfo<CL_PROGRAM_BINARIES>();
    std::cout << sizeof(output[0]) << std::endl;
    std::cout << output_sizes[0] << std::endl;

    const std::vector<unsigned long> binSizes = program.getInfo<CL_PROGRAM_BINARY_SIZES>();
    std::vector<char> binData (std::accumulate(binSizes.begin(),binSizes.end(),0));
    char* binChunk = &binData[0] ;


    //A list of pointers to the binary data    
    std::vector<char*> binaries;
    for(unsigned int i = 0; i<binSizes.size(); ++i) {
         binaries.push_back(binChunk) ;
         binChunk += binSizes[i] ;
    }

    program.getInfo(CL_PROGRAM_BINARIES , &binaries[0] ) ;
    std::ofstream binaryfile(writeFileName, std::ios::binary);
    for (unsigned int i = 0; i < binaries.size(); ++i)
        binaryfile.write(binaries[i], binSizes[i]);
    delete[] buffer;
    return program;
}
4
  • How did you create a binary from OpenCL C source code? Commented Dec 27, 2018 at 6:15
  • @AndrewSavonichev added it above Commented Dec 28, 2018 at 20:36
  • I'm not sure if this is a root cause, but you need to call cl::Program::build() even for a program created from binary. Commented Dec 28, 2018 at 21:30
  • @AndrewSavonichev that was it. It works now. I had no idea.. If you write it as the answer I'll tick it. BTW where in the docs is this mentioned. What is "building" if the program is already binary? Commented Dec 28, 2018 at 21:33

1 Answer 1

3

clBuildProgram must be called even for a program created by clCreateProgramWithBinary, according to the OpenCL Specification s5.6.2 "Building Program Executables":

OpenCL allows program executables to be built using the source or the binary. clBuildProgram must be called for program created using either clCreateProgramWithSource or clCreateProgramWithBinary to build the program executable for one or more devices associated with program. If program is created with clCreateProgramWithBinary, then the program binary must be an executable binary (not a compiled binary or library).

The reason for this is that a "device binary" is not necessary a fully compiled/linked machine code for a target device: it can be an intermediate representation in some form (e.g. LLVM IR), that requires further compilation.

Sign up to request clarification or add additional context in comments.

Comments

Your Answer

By clicking “Post Your Answer”, you agree to our terms of service and acknowledge you have read our privacy policy.

Start asking to get answers

Find the answer to your question by asking.

Ask question

Explore related questions

See similar questions with these tags.