Visible to Intel only — GUID: ixg1521225122200
Ixiasoft
Visible to Intel only — GUID: ixg1521225122200
Ixiasoft
5.5.5.3. Implementing Pipe Writes
Intel® only supports the convenience version of the write_pipe function. By default, write_pipe calls are nonblocking. Pipe write operations are successful only if there is capacity in the pipe to hold the incoming packet.
Where:
pipe_id identifies the buffer to which the pipe connects, and it must match the pipe_id of the corresponding read pipe (read_pipe).
data is the data that the pipe write operation writes to the pipe. It is a pointer to the packet type of the pipe. Note that writing to the pipe might lead to a global or local memory load, depending on the source address space of the data pointer.
<type> defines a pipe data width. The return value indicates whether the pipe write operation is successful. If successful, the return value is 0. If pipe write is unsuccessful, the return value is -1.
/*Declares the writable nonblocking pipe, p, which contains packets of type int*/
__kernel void kernel_write_pipe (__global const long *src,
write_only pipe int p)
{
for (int i = 0; i < N; i++)
{
//Performs the actual writing
//Emulates blocking behavior via the use of a while loop
while (write_pipe(p, &src[i]) < 0) { }
}
}
The while loop is unnecessary if you specify a blocking attribute. 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.