私はCUDAのモジュラ累乗のためのアプリケーションを開発しました。これは512ビット整数に対して非常によく機能します。この多倍精度整数は16個の32ビットワードに格納されます。私は2.5達成するために使用
いくつかの概念 - OpenSSLのべき乗剰余アプローチに比べ3.2高速化:CUDAの速度最適化
__shared__
メモリ- CUDAメモリは32ビット加算のため
- PTXコード、乗算
- を揃えます
整数は1024ビットに拡張しようとしていますが、パフォーマンスは劇的に低下します0.1から0.3までの整数であり、唯一の違いは整数を格納するのに必要なメモリサイズであり、今は32×32ビットワードである。数百倍も遅い2048ビットバージョンはもちろんですが、
たとえば、1000の累乗累乗(r = a^x mod n
)を計算する場合、すべてのオペランドをカーネルに送ります。これは512000バイトのメモリを意味します。
私の質問:この小さな変更がパフォーマンスに大きな影響を与えるのはなぜですか?
私はNvidia Geforce GT 520mx、Ubuntu 14.04 64ビットを使用します。
[mcve]がなければ、伝えるのは難しい/不可能です。あなたが見たいと思うかもしれない1つのポイントはレジスタ溢れです。しかし、これはちょうど推測です.... – havogt
登録の流出も私の最初の推測でした。 'nvcc -Xptxas = -v ...'でPTXをコンパイルすると、レジスタの使用状況に関する情報が得られます。 – Marco13
よろしくお願いいたします。 –