cuBLSA を用いて GPU で行列の積演算を行わせるにあたり、行列の積演算関数の引き数には、入力行列の転置の指定があり、転置によって計算時間にどう影響あるのか調査しました。
調査結果を公開します。

背景

深層学習の理解を深めようと自前で実装を行っていると、行列の積演算の所要時間を少しでも短くしようと感じます。　
BLASでの計算時間は、いかにメモリアクセスの時間を短縮するかの工夫（キャッシュメモリに当てる、メモリを連続的にアクセスする）の世界になっていると想像し、入力行列の転置によっても、所要時間に影響あるかと予想し、今回測定しました。

行列の積演算関数 gemm()

参照先「cuda tool kit documentation cublasgemm()」

gemm_sample.c

cublasSgemm(    
        handle,
        CUBLAS_OP_N, //行列A 転置有無
        CUBLAS_OP_N, //行列B 転置有無
        num,    // 行列Aの行数
        num,    // 行列Bの列数
        num,    // 行列Aの列数(=行列Ｂの行数)
        &alpha, // 行列の積に掛ける値(なければ1)
        devA,   // 行列A
        num,    // 行列Aの行数
        devB,   // 行列B
        num,    // 行列Bの行数
        &beta,  // 行列Cに掛けるスカラ値(なければ0)
        devC,   // 行列Cの初期値 兼 出力先
        num // 行列Cの行数
);

↑の行列A と行列B　の転置有無の引数を、

値	意味
CUBLAS_OP_N	転置なし
CUBLAS_OP_T	転置あり

の値を指定することで、転置の有無を指定します。

転置による所要時間

一辺 num の正方行列(下図)で、 num を増やしていった際の所要時間で比較する。

測定は「転置なし（無印）」「Aのみ転置(At)」「Bのみ転置(Bt)」「A,Bともに転置(AtBt)」の４種
測定は、下記グラフは、前回の記事「cuBLAS と cuBLAS-XT の調査。行列の積演算にて」で有力だった cuBLAS での測定。　cuBLAS-XTを含めたグラフは後述。
各num で１０回測定し、平均値をプロット。

↑を num 4096 以下で拡大したのが↓

↑を num 2048 以下で拡大したのが↓

結果

numの値に関わらず「Bのみ転置」が所要時間が最も短い。
「転置なし」は num が 2000 以下のあたりでは、「Bのみ転置」と同等。numが4096を超えたあたりから～3％位の差が生じる。

（「転置なし」と「Bのみ転置」間は予想より差が少なかった。「Aのみ転置」と「転置なし」と同様の差が、「転置なし」と「Bのみ転置」間にもあるように予想してました）

→「Bのみ転置」を用いるのが良さそう。

cuBLAS-XT も含めた比較

所要時間の短い方から、「cublas(Bのみ転置)」「cublasXt(Bのみ転置)」「cublas(転置なし)」「cublasXt(転置なし)」
「Bのみ転置」にしても、cuBLAS-XT より cuBLA のがわずかに所要時間が短い。（グラフィックカードが１枚の場合）

実装の話

BLASで扱う行列は Column優先(列優先)のメモリ配置。

画像などでよく用いる Row とは向きが異なるので注意が必要。

col優先での B転置！？　メモリにどう配置するん？

Row 優先表記で↓の配置。

特徴

MとPの共通項がどちらも行数になる。 (どっちがどっちでの悩みが一つ減る）
MとP、どちらをAにするかで、結果の転置を調整できる。

これはこれで便利そう。 (うまくできてるもんですね）

検算したの？

むしろ、結果から逆算しました...。

入力：
--- M ---
mem:[ 1.0000 ,2.0000 ,3.0000 ,4.0000 ,5.0000 ,6.0000 ,7.0000 ,8.0000 ,9.0000 ]
1.0000 2.0000 3.0000 
4.0000 5.0000 6.0000 
7.0000 8.0000 9.0000 

--- P ---
mem:[ 0.0001 ,0.0010 ,0.0100 ,0.1000 ,1.0000 ,10.0000 ,100.0000 ,1000.0000 ,10000.0000 ]
  0.0001    0.0010     0.0100 
  0.1000    1.0000    10.0000 
100.0000 1000.0000 10000.0000 

結果：
--- M×Pt ---
mem:[ 700.4001 ,800.5002 ,900.6003 ,7004.0010 ,8005.0020 ,9006.0030 ,70040.0100 ,80050.0200 ,90060.0300 ]
  700.4001   800.5002   900.6003 
 7004.0010  8005.0020  9006.0030 
70040.0100 80050.0200 90060.0300

--- P×Mt ---
mem:[700.4001 ,7004.0010 ,70040.0100 ,800.5002 ,8005.0020 ,80050.0200 ,900.6003 ,9006.0030 ,90060.0300 ]
700.4001 7004.0010 70040.0100 
800.5002 8005.0020 80050.0200 
900.6003 9006.0030 90060.0300

cuBLAS と cuBLAS-XT の調査（その２）。行列の積演算にて。転置の影響。

背景