たとえば、threadIdx.x を使用した複雑なインデックス計算がパフォーマンスに影響を与えるかどうかを自問しました。これらの変数は、カーネルがデバイスにアップロードされるとすぐに一定になりますか?
インデックスが threadIdx.x、threadIdx.y、および threadIdx.z に依存する巨大な配列に移動したいと考えています。たとえば、次のようなモジュロ演算が必要です
array[threadIdx.y % 2 + ...]
たとえば、threadIdx.x を使用した複雑なインデックス計算がパフォーマンスに影響を与えるかどうかを自問しました。これらの変数は、カーネルがデバイスにアップロードされるとすぐに一定になりますか?
インデックスが threadIdx.x、threadIdx.y、および threadIdx.z に依存する巨大な配列に移動したいと考えています。たとえば、次のようなモジュロ演算が必要です
array[threadIdx.y % 2 + ...]
インデックス計算に加算とモジュラスがあります。
CUDA プログラミング ガイドより: のスループットoperator+
は非常に高い (3.5 コンピューティング対応 GPU の場合は 160)。
operator%
に似たスループットで数十回の操作が必要operator+
です。
あなたの場合operator%
、リテラル定数を使用しており、コンパイラはそれを最適化する可能性が非常に高いです。また、定数は 2 のべき乗 (2) であるため、コンパイラはそれをビットごとに置き換えoperator&
ます (と同じスループットoperator+
)。
パフォーマンスを向上させずに算術演算を最適化するのに時間を浪費しないように、アプリケーションをプロファイリングすることが重要です。算術演算がメモリーのロードおよびストア操作によって完全に隠されていることはよくあることです。この場合、メモリー・スループットの最適化に集中する必要があります。
誰かが興味を持っている場合は、対応する PTX コードを評価しました。
(1) 複雑なスレッド ID の計算は、パフォーマンスに影響を与えます。「threadIdx.x」などは定数ではありません。
(2) "threadIdx.y % 2" は効率的に実装され、"threadIdx.y & 0x00000001" (Cuda Toolkit 5.5) に対応します。