OpenCLの：どのように[OK]をカーネル

Question

を適切にチェーンに、私は両方の入力と出力の画像を撮影し、いくつかの意味のある操作を行う2つのカーネルいる：今、私はチェーンにいくつかのケースでカーネルを

#pragma OPENCL EXTENSION cl_khr_3d_image_writes : enable 

kernel void Kernel1(read_only image3d_t input, write_only output) 
{ 
    //read voxel and some surrounding voxels 
    //perform some operation 
    //write voxel 
} 


#pragma OPENCL EXTENSION cl_khr_3d_image_writes : enable 

kernel void Kernel2(read_only image3d_t input, write_only output) 
{ 
    //read voxel and some surrounding voxels 
    //perform some other operation 
    //write voxel 
} 


#pragma OPENCL EXTENSION cl_khr_3d_image_writes : enable 

kernel void KernelCombined(read_only image3d_t input, write_only output) 
{ 
    //read voxel and some surrounding voxels 

    //... 
    //perform operation of both kernels (without read, write) 
    //... 
    //write voxel 
} 

をしたいです、だから私は最初にKernel 1を呼び出してからKernel2を呼び出すことができます。しかし、それは間に不必要な書き込みと読み込みがあることを意味します。私は両方を行う3番目のカーネルを書くこともできますが、コピー・ペースト・コードを維持することは面倒です。私は、私の知る限り、image3d_tの入力を渡すことができないので、実際には各カーネルの内容を別々の関数に入れることはできません。

質問： 2つのカーネルを繋ぐ巧妙な方法はありますか？おそらくOpenCLは私が知らない何か巧妙なことをしているのでしょうか？

編集：私がどのようにしたいのかの追加例。

Answer 1

あなたが求めていることを理解しています - イメージの書き込み/読み取りサイクルをカーネル間で削除したいと考えています。あなたが記述したカーネルでは、これは効率的ではありません。既存のカーネルでは、ボクセルとその周辺のボクセルを読み込みます。ボクセルを7つ読み込むことを意味します。カーネル2とカーネル3で同じ読み込みパターンを実行すると、計21回の読み込み（および3回の書き込み）が行われます。どうやらこれらの3つのカーネルを1つの出力ボクセルを書き込んだ単一のカーネルに連鎖すると、同じ結果を得るためにはもっと多くのソースボクセルから読み込む必要があります。

カーネルの書き込み/読み取りの連鎖が役立つシナリオは、カラーが隣接するものとは独立して変更される画像処理のようなシングルイン/シングルアウトカーネルの場合です。そのためには、あなたのカーネルに関するより高いレベルの記述と、あなたが持っている操作に基づいて必要なカーネルを生成できるものが必要です。

Answer 2

これは、opencl 2.0対応デバイスを使用している場合に可能です。 enqueue_kernelは、ホスト上のEnqueueNDRangeと同様に、カーネルが別のものをキューに入れることを許可します。

opencl 1.2を使用している場合、おそらく1.xのすべての場合、ホストに戻り、次のカーネルを呼び出す（または、次のカーネルがすでにキューに入れられている）必要があります。しかし、カーネル間でバッファーをホストにコピーする必要はありません。少なくとも、転送に何回も料金を払う必要はありません。

Answer 3

あなたの説明からわかる限り、特別なことはすべきではなく、OpenCL 1.2でもうまくいきます。

OpenCLコマンドキューはデフォルトでIN ORDERであり、カーネルコールの間にデータを転送する必要はありません。

デバイスにデータを残すだけで（マップ/マップ解除と読み取り/書き込みは行わないでください）、両方のカーネルをエンキューし、終了するまで待機します。ここではそれがどのように見えるかのコードスニペットは、次のとおりです。

// Enqueue first kernel 
clSetKernelArg(kernel1, 0, sizeof(cl_mem), in); 
clSetKernelArg(kernel1, 1, sizeof(cl_mem), out); 
clEnqueueNDRange(..., kernel1, ...); 

// Enqueue second kernel 
clSetKernelArg(kernel2, 0, sizeof(cl_mem), in); 
clSetKernelArg(kernel2, 1, sizeof(cl_mem), out); 
clEnqueueNDRange(..., kernel2, ...); 

// Flush the queue and wait for the results 
clFlush(...); // Start the execution 
clWait(...); // Wait until all operations in the queue are done 

OOO（ORDER OF OUT）1は、カーネル間の依存関係を指定するイベント（clEnqueueNDRangeKernelの最後の3つのparamsを参照）を使用してのclWaitForEventsを行うことができますキューを使用してあなたのパイプラインの終わり。