OpenCL* for CPU
Ask questions and share information on Intel® SDK for OpenCL™ Applications and OpenCL™ implementations for Intel® CPU
Announcements
This forum covers OpenCL* for CPU only. OpenCL* for GPU questions can be asked in the GPU Compute Software forum. Intel® FPGA SDK for OpenCL™ questions can be ask in the FPGA Intel® High Level Design forum.
1663 Discussions

How to use Pipe in OpenCL2.0

lucas_w_2
Beginner
307 Views

Hi guys!
 i am tring to use Pipe in OpenCL2.0. But I am confused about the built-in Pipe read and write functions,such as write_pipe, reserver_read_pipe.

Code :
reserve_id_t reserve_write_pipe (
                                        pipe gentype p,
                                       uint num_packets)
Code :
int write_pipe (pipe gentype p,
                                reserve_id_t reserve_id,
                                uint index,
                               const gentype *p


1. For the reserve_write_pipe() function, Description is that 

Code :
Reserve num_packets entries for writing to pipe p. Returns a valid reservation ID if the reservation is successful.

I don't know what's the meaning of the num_packets argument.
2. For write_pipe() function,Description is that 

Code :
Write packet specified by ptr to the reserved area of the pipe referred to by reserve_id and index.

What's the meaning of "reserved area of the pipe"? and What's the meaning of "reserve_id " and "index"?
3. If i want to write to Pipe by order(such as work-item 0 writes 0 to the pipe,work-items 1 write 1 to the pipe,.. work-item n writes n to the pipe,so the data in the pipe is "012345……n" ,and 0 is the first one in the pipe),what should I do ?
Can someone help me please? Thanks.
hi_buddy

0 Kudos
6 Replies
Robert_I_Intel
Employee
307 Views

I would highly discourage you to use pipes, but if you insist:

1. Let's say you have an int pipe and num_packets is 5: you reserve 5 int slots in the reservation id that is returned (let's say 42), so then you can write to these slots by providing an index (0 thru 4) and your reservation id (42).

2. What you can do is the first work item in a work group makes a reservation the size of the work group. Then every work item writes to the slot by providing its local id as an index.

Just curious: why do you want to use pipes? Do you actually have an algorithm that you think will map to pipes or just want to press all the buttons in OpenCL 2.0 spec?

lucas_w_2
Beginner
307 Views

Hi Robert. I just want to have a try , because Pipe is the new feature in OpenCL2.0

2. What you can do is the first work item in a work group makes a reservation the size of the work group. Then every work item writes to the slot by providing its local id as an index.

I don't understand what have you said above. Can you explain that ? thanks a lot

 

   

Robert_I_Intel
Employee
307 Views
local reserve_id_t rid;
if (get_local_id(0) == 0)
    rid = reserve_write_pipe (p, get_local_size(0));

barrier(CLK_LOCAL_MEM_FENCE);
int foo = get_local_id(0);
write_pipe (p, rid, get_local_id(0), &foo);

work_group_commit_write_pipe(p, rid);

 

 

 

Robert_I_Intel
Employee
307 Views

You can even make it simpler:

reserve_id_t rid = work_group_reserve_write_pipe(p, get_local_size(0));
int foo = get_local_id(0); 
write_pipe (p, rid, get_local_id(0), &foo); 
work_group_commit_write_pipe(p, rid); 

 

lucas_w_2
Beginner
307 Views

Hi Robert,

1. I have modified the code under your guidance, now  I can get the sequential data across a workgroup,not across the NDRange. it seems that there is no way to control the order across the NDRange.

2.I'm curious that why you highly discourage me to use pipes.

Thanks.

 

Robert_I_Intel
Employee
307 Views

Hi Lucas,

1. Correct: work groups execute independently, so you won't be able to control the order across the NDRange.

2. Pipes are a weird beast in OpenCL: usually, when you talk pipes, concurrently executing processes come to mind. Typically, people expect kernels to run concurrently and communicate via pipes: that's not what is going on here. Pipes are just glorified buffers - there is no special magic and anything you can accomplish with pipes you can do with buffers. There is no concurrency either: one kernel completes execution and then the other starts, so you need to size the pipe so it could fit ALL the output of the kernel - it is just a memory buffer :(

Reply