インテルのみ表示可能 — GUID: ewa1425924338976
Ixiasoft
インテルのみ表示可能 — GUID: ewa1425924338976
Ixiasoft
5.5.5.5. depth属性を使用するバッファーされたパイプの実装
バッファーされたパイプを使用し、スループットの制限や共有メモリーのアクセスの同期化といったデータのトラフィックを制御できます。バッファーされていないパイプにおいて書き込み動作は、読み出し動作がデータの読み取りをしようとしている場合にのみ開始できます。バッファーされていないパイプは、並行して実行されるカーネルでのブロッキングの読み書き動作と組み合わせて使用してください。バッファーされていないパイプは、自己同期型のデータ転送を効率的に提供します。
バッファーされたパイプにおいて書き込み動作は、受信するパケットを収容する容量がパイプにある場合にのみ進めることが可能です。読み出し動作は、少なくとも1つのパケットがパイプになければ実行することができません。
パイプ呼び出しが書き込みカーネルと読み出しカーネルで異なって表されている場合にバッファーされたパイプを使用すると、カーネルは並行して実行されません。
__kernel void producer (__global int *in_data, write_only pipe int __attribute__((blocking)) __attribute__((depth(10))) c) { for (i = 0; i < N; i++) { if (in_data[i]) { write_pipe( c, &in_data[i] ); } } } __kernel void consumer (__global int *check_data, __global int *out_data, read_only pipe int __attribute__((blocking)) c ) { int last_val = 0; for (i = 0; i < N; i++) { if (check_data[i]) { read_pipe( c, &last_val ); } out_data[i] = last_val; } }
この例において書き込み動作は、10個のデータ値をパイプに正常に書き込むことができます。パイプがフルの状態になると書き込みカーネルは、読み出しカーネルがパイプのデータの一部を消費するまで失敗を返します。
パイプの読み出しと書き込みの呼び出しは条件付きステートメントのため、パイプの読み出しおよび書き込みの呼び出しには不均衡が発生する可能性があります。パイプにバッファー容量を追加すると、producerとconsumerカーネルを切り離すことができます。この方法は、consumerカーネルがパイプからデータを読み取っていない際に、producerカーネルがパイプにデータを書き込んでいる場合に特に重要です。