1 Reply Latest reply on Nov 21, 2011 7:47 PM by notzed

    Error:clBuildprogram(-11)

    meetajinkya88
      cant understand the error

      I have 64 bit fedora with ATI Radeon GPU.I am doing my project in Image processing using OpenCL .since I am new in this field,while execution following error comes:

      Error:clBuildprogram(-11)

      what does that error mean?

      I am executing sample code of 1-D array.It works fine but while handling with images data it gives this error.So plz help me to get rid of this error.

      Following is my code:

       

      #include <iostream> #include <cstdlib> #include <fstream> #include <string> #include <stdio.h> #include <stdlib.h> #include "fitsio.h" #ifdef __APPLE__ #include <OpenCL/opencl.h> #else #include <CL/cl.h> #endif //#define MAT_SIZE 262144 using namespace std; class FitsFileInfo{ public: int bitpix; int naxis; long naxes[2]; }; void mergesort(float a[], long long int low, long long int high); void err_check( int err, string err_code ) { if ( err != CL_SUCCESS ) { cout << "Error: " << err_code << "(" << err << ")" << endl; exit(-1); } } //****************************** Read Fits File *********************************** int readfits(float **a, char *fname, FitsFileInfo *ffinfo) { long abc=0; fitsfile *fptr; // FITS file pointer, defined in fitsio.h int status = 0; /* CFITSIO status value MUST be initialized to zero! */ int bitpix, naxis, ii, anynul; long naxes[2] = {1,1}, fpixel[2] = {1,1}; double *pixels; char format[20], hdformat[20]; long MAT_SIZE; if (!fits_open_file(&fptr,fname, READONLY, &status)) { if (!fits_get_img_param(fptr, 2, &bitpix, &naxis, naxes, &status) ) { cout << "bitpix = " << bitpix << " naxis = " << naxis << " naxes = [ " << naxes[0] << ", " << naxes[1] << " ]" << endl; MAT_SIZE = naxes[0] * naxes[1]; ffinfo->bitpix = bitpix; ffinfo->naxis = naxis; ffinfo->naxes[0] = naxes[0]; ffinfo->naxes[1] = naxes[1]; *a = (float *)malloc(MAT_SIZE * sizeof(float)); float *mat_a; mat_a = *a; if (naxis > 2 || naxis == 0) printf("Error: only 1D or 2D images are supported\n"); else { // get memory for 1 row pixels = (double *) malloc(naxes[0] * sizeof(double)); if (pixels == NULL) { printf("Memory allocation error\n"); return(1); } // loop over all the rows in the image, top to bottom for (fpixel[1] = naxes[1]; fpixel[1] >= 1; fpixel[1]--) { if (fits_read_pix(fptr, TDOUBLE, fpixel, naxes[0], NULL,pixels, NULL, &status) ) // read row of pixels break; /// jump out of loop on error for (ii = 0; ii < naxes[0]; ii++) { *(mat_a + abc) =(float) pixels[ii]; abc++; } } free(pixels); } } cout<<endl<<"Matrix Size = "<<MAT_SIZE<<endl; fits_close_file(fptr, &status); } } //**************************** Write Data into Fits File *********************************** int writefits(float *data,FitsFileInfo *ffinfo ) { long fpixel[]={1,1}, MAT_SIZE; fitsfile *ofptr; int status = 0; float tmp; long siz = ffinfo->naxes[1]; for(long i=0, ii = (ffinfo->naxes[0]-1); i<ii; ++i, --ii) for(long j=0; j<ffinfo->naxes[1]; ++j) { tmp = data[ (i*siz) +j]; data[ (i*siz)+j] = data[ (ii*siz) +j]; data[ (ii*siz) +j] = tmp; } MAT_SIZE = ffinfo->naxes[0] * ffinfo->naxes[1]; fits_create_file(&ofptr, "skysub.fits", &status); fits_create_img(ofptr, ffinfo->bitpix, ffinfo->naxis, ffinfo->naxes, &status); fits_write_pix(ofptr, TFLOAT, fpixel,MAT_SIZE, data, &status ); fits_close_file(ofptr, &status); return(0); } //********************************************************************************************* int main() { cl_platform_id platform_id = NULL; cl_device_id device_id = NULL; cl_context context = NULL; cl_command_queue command_queue = NULL; cl_mem mobj_A = NULL, mobj_D; cl_program program = NULL; cl_kernel kernel = NULL; cl_uint ret_num_devices; cl_uint ret_num_platforms; cl_int err; float *mat_a; char fname[]= "a512.fits"; FitsFileInfo ffinfo; long MAT_SIZE; float ssum; readfits(&mat_a, fname, &ffinfo); MAT_SIZE = ffinfo.naxes[0] * ffinfo.naxes[1]; ssum = 0; cout<< MAT_SIZE; /* for ( cl_int i = (MAT_SIZE - 256); i < MAT_SIZE; i++ ) { // ssum += mat_a[i]; // if (mat_a[i] > 0 && mat_a[i]<99.6) cout << mat_a[i]<<" "; } ssum /= MAT_SIZE; cout<< "Ssum = "<< ssum; // int iii; // cin>> iii; */ // Step 01: Get platform/device information err = clGetPlatformIDs( 1, &platform_id, &ret_num_platforms ); err_check( err, "clGetPlatformIDs" ); // Step 02: Get information about the device err = clGetDeviceIDs( platform_id, CL_DEVICE_TYPE_GPU, 1, &device_id, &ret_num_devices ); err_check( err, "clGetDeviceIDs" ); // Step 03: Create OpenCL Context context = clCreateContext( NULL, 1, &device_id, NULL, NULL, &err ); err_check( err, "clCreateContext" ); // Step 04: Create Command Queue command_queue = clCreateCommandQueue( context, device_id, 0, &err ); err_check( err, "clCreateCommandQueue" ); // Step 05: Create memory objects and tranfer the data to memory buffer mobj_A = clCreateBuffer( context, CL_MEM_READ_WRITE, MAT_SIZE * sizeof(float), NULL, &err ); mobj_D = clCreateBuffer( context, CL_MEM_READ_WRITE, MAT_SIZE / 512 * sizeof(float), NULL, &err ); err = clEnqueueWriteBuffer( command_queue, mobj_A, CL_TRUE, 0, MAT_SIZE * sizeof(float), mat_a, 0, NULL, NULL ); err_check( err, "clEnqueueWriteBuffer" ); // Step 06: Read kernel file ifstream file("par_sum_sort.cl"); string prog( istreambuf_iterator<char>( file ), ( istreambuf_iterator<char>() ) ); const char *source_str = prog.c_str(); // Step 07: Create Kernel program from the read in source program = clCreateProgramWithSource( context, 1, (const char **) &source_str, 0, &err ); err_check( err, "clCreateProgramWithSource" ); // Step 08: Build Kernel Program err = clBuildProgram( program,1, &device_id, NULL, NULL, NULL ); err_check( err, "clBuildProgram" ); // Step 09: Create OpenCL Kernel kernel = clCreateKernel( program, "sum", &err ); err_check( err, "clCreateKernel" ); // Step 10: Set OpenCL kernel argument err = clSetKernelArg( kernel, 0, sizeof( cl_mem ), (void *) &mobj_A ); err_check( err, "clSetKernelArg 0" ); err = clSetKernelArg( kernel, 1, sizeof( cl_mem ), (void *) &mobj_D ); err_check( err, "clSetKernelArg 1" ); err = clSetKernelArg( kernel, 2, 512 * sizeof( cl_float ), NULL ); err_check( err, "clSetKernelArg 2" ); // Step 11: Execute OpenCL kernel in data parallel size_t worksize[] = { MAT_SIZE, 1, 1 }; clEnqueueNDRangeKernel( command_queue, kernel, 1, NULL, worksize, 0, 0, 0, 0 ); err_check( err, "clEnqueueNDRangeKernel" ); // Step 12: Read (Transfer result) from the memory buffer //------------------------ long twsize; twsize = worksize[0]/512; cl_event event[2]; int ii = 0; while (twsize > 0) { err = clSetKernelArg( kernel, 0, sizeof( cl_mem ), (void *) &mobj_D ); err_check( err, "clSetKernelArg 3" ); err = clSetKernelArg( kernel, 1, sizeof( cl_mem ), (void *) &mobj_D ); err_check( err, "clSetKernelArg 4" ); err = clSetKernelArg( kernel, 2, 512 * sizeof( cl_float ), NULL ); err_check( err, "clSetKernelArg 5" ); worksize[0] = worksize[0] / 512; clEnqueueNDRangeKernel( command_queue, kernel, 1, NULL, worksize, 0, 0, 0,&event[(ii%2)] ); err_check( err, "clEnqueueNDRangeKernel" ); err = clWaitForEvents(1, &event[(ii%2)]); err_check( err, "clWaitForEvents" ); ii = ii+1; twsize = twsize / 512; } //********************************************************************************************************** /* //------------- Sorting Part -------------------- cl_kernel kernel1; kernel1 = clCreateKernel( program, "bitonic_sort", &err ); err_check( err, "clCreateKernel" ); cl_int N; cl_event event1[2]; for (long i=0;i<MAT_SIZE; ++i) { N = i % 2; // Step 10: Set OpenCL kernel argument err = clSetKernelArg( kernel1, 0, sizeof( cl_mem ), (void *) &mobj_A ); err_check( err, "clSetKernelArg 0" ); err = clSetKernelArg( kernel1, 1, sizeof( cl_int ),(void *)&N ); err_check( err, "clSetKernelArg 1" ); // Step 11: Execute OpenCL kernel in data parallel worksize[0] = MAT_SIZE; clEnqueueNDRangeKernel( command_queue, kernel1, 1, NULL, worksize, 0, 0, 0, &event1[ (i%2) ] ); err_check( err, "clEnqueueNDRangeKernel" ); err = clWaitForEvents(1, &event1[(i%2)]); err_check( err, "clWaitForEvents" ); } */ //********************************************************************************************************** float result[MAT_SIZE/512]; for( int i=0; i< (MAT_SIZE/512) ; ++i) result[i] = 0; err = clEnqueueReadBuffer( command_queue, mobj_D, CL_TRUE, 0, MAT_SIZE / 512 * sizeof(float), result, 0, NULL, NULL ); float sdata[MAT_SIZE]; // err = clEnqueueReadBuffer( command_queue, mobj_A, CL_TRUE, 0, MAT_SIZE* sizeof(float), sdata, 0, NULL, NULL ); for(int i=0; i<MAT_SIZE;i++) sdata[i] = *(mat_a +i); mergesort(sdata,0,MAT_SIZE); // Display result cout<<"Partial sum = "<<result[1]; cout<<"sum = "<<result[0]<<endl; float mean_val; // mean_val = (result[0]/MAT_SIZE); mean_val = result[0]; cout<<"Mean = "<<mean_val<<endl; /* for (cl_int i = 0; i<MAT_SIZE; ++i) { cout<< sdata[i] << " "; } */ err = clReleaseMemObject( mobj_A ); err = clReleaseMemObject( mobj_D ); float med_val; med_val = sdata[MAT_SIZE/2]; // med_val = 99.89; cout<<endl; cout<<"Median Value = "<<med_val<<endl<<endl; float mode_val; mode_val = ( (2.5 * med_val) - (1.5 * mean_val) ); // mode_val = ( (3 * med_val) - (2 * mean_val) ); cout<<"Mode Value = "<<mode_val<<endl<<endl; // mode_val = 110; //****************** Image Thresholding and writing data back in fits formate *************************** mobj_A = clCreateBuffer( context, CL_MEM_READ_WRITE, MAT_SIZE * sizeof(float), NULL, &err ); err_check( err, "Thresholding :clCreateBuffer" ); err = clEnqueueWriteBuffer( command_queue, mobj_A, CL_TRUE, 0, MAT_SIZE * sizeof(float), mat_a, 0, NULL, NULL ); err_check( err, "Thresholding :clEnqueueWriteBuffer" ); cl_kernel kernel2; kernel2 = clCreateKernel( program, "thresholding", &err ); err_check( err, "clCreateKernel 2 : Thresholding " ); cl_event event2; err = clSetKernelArg( kernel2, 0, sizeof( cl_mem ), (void *) &mobj_A ); err_check( err, "Thresholding : clSetKernelArg 0 " ); err = clSetKernelArg( kernel2, 1, sizeof( cl_float ),(void *)&mode_val ); err_check( err, "Thresholding : clSetKernelArg 1" ); worksize[0] = MAT_SIZE; clEnqueueNDRangeKernel( command_queue, kernel2, 1, NULL, worksize, 0, 0, 0, &event2 ); err_check( err, "clEnqueueNDRangeKernel for Thresholding " ); // err = clWaitForEvents(1, &event2); err_check( err, "clWaitForEvents : event 2 " ); //float tdata[MAT_SIZE]; err = clEnqueueReadBuffer( command_queue, mobj_A, CL_TRUE, 0, MAT_SIZE* sizeof(float), mat_a, 0, NULL, NULL ); double chk1; chk1 = 0; /* for (cl_int i = 0; i<MAT_SIZE; ++i) { if (chk1 < tdata[i]) chk1 = tdata[i]; //cout<< tdata[i] << " "; } cout<<"max value = "<< chk1; */ writefits(mat_a, &ffinfo); //********************************************************************************************************** // Step 13: Free objects err = clFlush( command_queue ); err = clFinish( command_queue ); err = clReleaseKernel( kernel ); // err = clReleaseKernel( kernel1 ); err = clReleaseKernel( kernel2 ); err = clReleaseProgram( program ); err = clReleaseMemObject( mobj_A ); // err = clReleaseMemObject( mobj_D ); err = clReleaseCommandQueue( command_queue ); err = clReleaseContext( context ); return 0; } Kernel code: __kernel void sum( __global double *a, __global double *b, __local double *c ) { int gid = get_global_id(0); int bid = get_group_id(0); int ls = get_local_size(0); int lid = get_local_id(0); c[lid] = a[gid]; barrier(CLK_LOCAL_MEM_FENCE); for( int s = (ls/2); s>0; s >>= 1) { if( lid < s) { c[lid] = ( c[lid] + c[lid + s] ) / 2; } barrier(CLK_LOCAL_MEM_FENCE); } if(lid == 0) b[bid] = c[0]; } //*************************** Sorting Part ********************************* __kernel void bitonic_sort( __global float *a, int i) { int gs = get_global_size(0); int gid = get_global_id(0); if( (i == 0) && (gid % 2) == 0 && ((gid + 1) < gs) && a[gid] > a[gid+1]) { a[gid] += a[gid + 1]; a[gid +1] = a[gid] - a[gid +1]; a[gid] -= a[gid + 1]; } if( (i == 1) && (gid % 2) != 0 && ((gid + 1) < gs) && a[gid] > a[gid+1]) { a[gid] += a[gid + 1]; a[gid +1] = a[gid] - a[gid +1]; a[gid] -= a[gid + 1]; } barrier(CLK_LOCAL_MEM_FENCE); } //*************************** Thresholding Kernel Fun ********************* __kernel void thresholding(__global float *a, float mode_val) { int gid = get_global_id(0); if( a[gid] - mode_val >= 0) a[gid] = a[gid] - mode_val; else a[gid] = 0; }