3 Replies Latest reply on Jan 31, 2017 10:53 AM by russdirks

    clBuildProgram error :  The binary is incorrect or incomplete.


      I'm trying to load and run a binary kernel that was compiled offline.  clBuildProgram is failing with CL_BUILD_PROGRAM_FAILURE (-11), and when I dump the build log, I get the following:


      Build log:
      Error: The binary is incorrect or incomplete. Finalization to ISA couldn't be performed.

      clBuildProgram failed


      My system is Windows 10, Radeon 17.1.1 running a R9 380X.


      This is the code I used to export the binary:


      void exportBinary(std::string filename)
        cl_uint nDevices = 0;
        program.getInfo(CL_PROGRAM_NUM_DEVICES, &nDevices);
        size_t *pgBinarySizes = new size_t[nDevices];
        program.getInfo(CL_PROGRAM_BINARY_SIZES, pgBinarySizes);
        unsigned char **pgBinaries = new unsigned char*[nDevices];
        for (cl_uint i = 0; i < nDevices; i++)
        pgBinaries[i] = new unsigned char[pgBinarySizes[i]];
        program.getInfo(CL_PROGRAM_BINARIES, pgBinaries);
        std::fstream f(filename.c_str(), (std::fstream::out | std::fstream::binary));
        f.write((const char*) pgBinaries[0], pgBinarySizes[0]);



      This is the code I used to load the binary:


      #define MAX_BINARY_SIZE (0x100000)  
      fp = fopen(fileName, "r");      
      binary_buf = (char *) malloc(MAX_BINARY_SIZE);      
      binary_size = fread(binary_buf, 1, MAX_BINARY_SIZE, fp);      
      ret = clGetPlatformIDs(1, &platform_id, &ret_num_platforms);      
      ret = clGetDeviceIDs(platform_id, CL_DEVICE_TYPE_DEFAULT, 1, &device_id, &ret_num_devices);  
      /* Create OpenCL context*/      
      context = clCreateContext(NULL, 1, &device_id, NULL, NULL, &ret); 
      /* Create kernel program from the kernel binary */      
      program = clCreateProgramWithBinary(context, 1, &device_id, (const size_t *) &binary_size,           
        (const unsigned char **) &binary_buf, &binary_status, &ret);       
      ret = clBuildProgram(program, 1, &device_id, NULL, NULL, NULL);      
      if (ret != CL_SUCCESS)      
        dumpBuildLogFile(program, device_id);           
        return false;      


      The kernel is just a very simple test program that adds a number to itself:


      __kernel void vecAdd(__global float* a) 
           int gid = get_global_id(0);     
            a[gid] += a[gid];  


      Hopefully someone can give me a clue as to what is going wrong.


      PS: this is my third post on these forums.  Hopefully I can be white listed soon.