2012-05-15 7 views
6

として枝を書き換えます非分岐のスタイルは:CUDA/openCL;コードは常に(あるいは、少なくとも、ほとんどの時間が)に書き換えることができ</p> <pre><code>for (int i=0; i<width; i++) { if(i % threadIdx.x == 0) quantity += i*i; } </code></pre> <p>:のようにブランチが、CUDAやOpenCLのプログラムで必要とされる時間のほとんどは非分岐表現

for (int i=0; i<width; i++) 
{ 
    quantity += i*i* (i % threadIdx.x != 0); 
} 

トレードオフはちょうどその時々の値である、いずれかの第二の場合には(すべてのスレッド上でより多くの計算をやって対単一ワープスロットで実行して、合計が常に実行されているように見えますゼロ)

分岐操作が可能な分岐ごとに複数のワープ・スロットを取ると仮定すると、第2のものは第1のものよりも一貫して優れていると思われます。コンパイラに頼って、2)理に適ったときに、あるいは広範に適用できる基準がないときに最適化することができますか?一般に、試したりプロファイリングすることなくどちらが良いかを決めることはできません。

+0

幅はどのくらいですか?幅がかなり大きいことがわかっている場合は、どの値を使用するか分かるので、forループを繰り返してはいけません。 'While(i 3Pi

答えて

0

私はCUDAについての多くの思い出を持っていませんが、なぜループを並列化しないのですか?演算を追加するには、アトミック演算[1]を使用する必要があります。これがあなたを助けることを願っています!申し訳ありませんが、そうでない場合。

  1. アトミックオペレーション:私の経験でhttp://supercomputingblog.com/cuda/cuda-tutorial-4-atomic-operations/
+0

これはカーネル内部にあり、各スレッドは完全なループを実行しています。コメントは意味をなさない。 –

1

- それは完全にコンパイラの作家までのエッジ例これらの種類を最適化することです。

1)を2)にすることはできません。ここでは1つです:私は10個のスレッドごとに計算の特定の部分を実行する方が効率的だったカーネルを書いています。そのような最適化は推論できませんが(数式を減算する)条件付き対「すべて実行」の結果にかかわらず同じ結果が得られます。

しかし、threadId == 0のチェックが一般的なシナリオであっても、実際に最適化されているかどうかはわかりません。私は実装に依存し、デバイス自体(CPU対GPU)にも依存していると賭けるだろう。

上記の理由だけでなく、作業スケジューラがスレッドのセットを開始/停止するのにどれだけ費用がかかっているかによって動作が異なる可能性があるため、最も効果的なものを実際に見つけるために試してみる必要がありますそれらがすべて実行される(ほとんどの場合、ゼロ/アイデンティティの結果を提供する)こととは対照的です。

希望すると便利です。

+0

あなたの経験上、最悪のシナリオを想定して、私がいつもスタイル2でコードを書くべきかどうかについて、いくつかの勧告をすることができますか?意図しない結果が生じることがありますか? – lurscher

+0

私はすべての場合において、一方を他方よりも正当化することはできません - それが私の要点です。おそらく1)CPUデバイス上で扱えるものを減らすようなことをしていたら、2)古いハードウェアで分岐するコストのためにGPUを使用していたのです。考慮すべき要因は、デバイスの種類、非並列の計算方法、計算を複数のカーネル(おそらく削減)に分割すること、最後に選択されたデバイスタイプのすべてのハードウェアの分岐オーバーヘッドが受け入れられるかどうかです。しかし、IMO、実験は常に推奨されるでしょう。 – Ani

+0

GPUデバイスの特定のケースで話していますが、ブランチ予測とパイプライニングがたくさんあり、レイテンシ隠蔽に役立っているので、明らかにCPUの利益はありません – lurscher

3

モジュロ演算はかなり高価です:モジュロを加算すると、スレッドが1つしか実行されないという単一の命令を持つよりも時間がかかることは合理的に確信しています。 1つの分岐ステートメント(ifelseがない場合)は、他のスレッドだけがハングしますが、ステートメントが実行されている場合は、スレッドがハングします。 gpusは非常に高速なコンテキスト切り替え用に最適化されているため、そのためのコストはほとんどありません。

長い分岐文を使用しないことをお勧めします.GPU上の並列計算が多すぎると(つまり、すべての作業を行うスレッドが1つ)、並列処理の利点が無効になります。

+0

また、CUDA Best Programming Guideをチェックするだけで、分岐予測にコードを使いやすくすることが優先されます。一般的に最適化する重要な要素があります。 – 3Pi

関連する問題

 関連する問題