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

    clBuildProgram error :  The binary is incorrect or incomplete.

    russdirks

      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]);
        f.close();
      }
      

       

       

      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);      
      fclose(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.