Visible to Intel only — GUID: ibv1521225127778
Ixiasoft
Visible to Intel only — GUID: ibv1521225127778
Ixiasoft
5.5.5.4. Implementing Pipe Reads
Intel® only supports the convenience version of the read_pipe function. By default, read_pipe calls are nonblocking.
Where:
pipe_id identifies the buffer to which the pipe connects, and it must match the pipe_id of the corresponding pipe write operation (write_pipe).
data is the data that the pipe read operation reads from the pipe. It is a pointer to the location of the data. Note that write_pipe call might lead to a global or local memory load, depending on the source address space of the data pointer.
<type> defines the packet size of the data.
/*Declares the read_only_pipe that contains packets
of type long.*/
/*Declares that read_pipe calls within the kernel will exhibit
blocking behavior*/
__kernel void kernel_read_pipe(__global long *dst,
read_only pipe long __attribute__((blocking)) p)
{
for (int i = 0; i < N; i++)
{
/*Reads from a long from the pipe and stores it
into global memory at the specified location*/
read_pipe(p, &dst[i]);
}
}
To facilitate better hardware implementations, Intel® provides facility for blocking write_pipe calls by specifying the blocking attribute (that is, __attribute__((blocking))) on the pipe argument declaration for the kernel. Blocking write_pipe calls always return success.