Build error when using pipe related functions in openCL code

Question

I am writing to use the pipe feature of opencl 2.0. I have a kernel which filters the input and writes into a pipe. But the issue is that when i use pipe related functions i get the error "function declared implicitly". When i checked online i found that this error usualy appears in c code if the function is not declared before its used. But this is an openCL library function.

My kernel code is:

__kernel void filter1_kernel(__global int *input, const unsigned int rLen, const unsigned int lower, const unsigned int upper, __global int *pipe1){

unsigned int bitmap = 0;
int count = 0;
//reserve_id_t rid;

uint numWorkItems = get_global_size(0);
uint tid          = get_global_id(0);
uint num_packets = get_local_size(0);

while(tid < rLen){
    if(input[tid] >= lower && input[tid] <= upper){
        bitmap = bitmap | (1 << count);
        printf((__constant char *)"\nfilter1 %u %u %d %u", tid, bitmap, input[tid], count);
    }
    tid += numWorkItems;
    count++;
}

reserve_id_t rid = work_group_reserve_write_pipe(pipe1, num_packets);
while(is_valid_reserve_id(rid) == false) {
    rid = work_group_reserve_write_pipe(pipe1, num_packets);
}
//write to pipe

}

The errors that i got are:

Build Log Buffer length : 1048 --- Build log --- "C:\Users\pdca\AppData\Local\Temp\OCLFB5E.tmp.cl", line 40: error: function "work_group_reserve_write_pipe" declared implicitly reserve_id_t rid = work_group_reserve_write_pipe(pipe1, num_packets); ^

"C:\Users\pdca\AppData\Local\Temp\OCLFB5E.tmp.cl", line 40: warning: a value of type "int" cannot be used to initialize an entity of type "reserve_id_t" reserve_id_t rid = work_group_reserve_write_pipe(pipe1, num_packets); ^

"C:\Users\pdca\AppData\Local\Temp\OCLFB5E.tmp.cl", line 41: error: function "is_valid_reserve_id" declared implicitly while(is_valid_reserve_id(rid) == false) { ^

"C:\Users\pdca\AppData\Local\Temp\OCLFB5E.tmp.cl", line 42: warning: a value of type "int" cannot be assigned to an entity of type "reserve_id_t" rid = work_group_reserve_write_pipe(pipe1, num_packets); ^

2 errors detected in the compilation of "C:\Users\pdca\AppData\Local\Temp\OCLFB5 E.tmp.cl". Frontend phase failed compilation.

                            --- Build log ---

ERROR: clBuildProgram (CL_BUILD_PROGRAM_FAILURE)

ThorngardSO ThorngardSO · Accepted Answer · 2016-02-06T11:49:01

From the CL-specs (https://www.khronos.org/registry/cl/specs/opencl-2.0.pdf), page 203:

If the cl-std build option is not specified, the highest OpenCL C 1.x language version supported by each device is used when compiling the program for each device. Applications are required to specify the –cl-std=CL2.0 option if they want to compile or build their programs with OpenCL C 2.0.

So if you did not include this option to your clBuildProgram() call, the CL-compiler will not recognize any 2.0 language-features. Thus, your call should look sth like this:

clBuildProgram (program, num_devices, device_list, "–cl-std=CL2.0", NULL, NULL);

In addition, I think your kernel-parameter is not correct. You cannot use a __global int *pipe1 as an argument for the pipe functions. It should probably be declared as __write_only pipe int pipe1.

Build error when using pipe related functions in openCL code

1 Answers