Main Content

gpucoder.stridedMatrixMultiplyAdd

加算を伴うストライド バッチ行列乗算演算の最適化された GPU 実装

ページ内をすべて折りたたむ

構文

D = gpucoder.stridedMatrixMultiplyAdd(A,B,C)
___ = gpucoder.stridedMatrixMultiplyAdd(___,Name,Value)

説明

D = gpucoder.stridedMatrixMultiplyAdd(A,B,C) は、行列のバッチについてストライド行列-行列乗算と加算を実行します。バッチの各インスタンスの入力行列 A、B、および C は、前のインスタンスのアドレスと同じアドレス オフセットに位置します。関数 gpucoder.stridedMatrixMultiplyAdd は次の形式の行列-行列乗算を実行します。

D=αAB+βC

ここで、α および β はスカラー乗算係数、A、B、C、および D はそれぞれ m 行 k 列、k 行 n 列、m 行 n 列、および m 行 n 列の次元をもつ行列です。A および B は、転置またはエルミート共役にすることもできます。既定の設定では、α および β は 1 に設定され、行列は転置されません。さまざまなスカラー乗算係数を指定し、入力行列の転置演算を実行するには、Name,Value ペアの引数を使用します。

関数 gpucoder.stridedMatrixMultiplyAdd に渡されるすべてのバッチは一様でなければなりません。つまり、すべてのインスタンスは同じ次元 m,n,k でなければなりません。

例

___ = gpucoder.stridedMatrixMultiplyAdd(___,Name,Value) は、1 つ以上の Name,Value ペアの引数で指定されたオプションを使用することにより、バッチ行列乗算演算と加算演算を実行します。

例

すべて折りたたむ

加算を伴うシンプルなバッチ行列-行列乗算を実行し、関数 gpucoder.stridedMatrixMultiplyAdd を使用して、適切な cublas<t>gemmStridedBatched API を呼び出す CUDA® コードを生成します。

1 つのファイル内に、行列入力 A、B、および C を受け入れるエントリポイント関数 myStridedMatMulAdd を記述します。入力行列は転置しないため、'nn' オプションを使用します。 

function [D] = myStridedMatMulAdd(A,B,C,alpha,beta)

[D] = gpucoder.stridedMatrixMultiplyAdd(A,B,C,'alpha',alpha,...
    'beta',beta,'transpose','nn');

end

コード生成で使用する double の行列の型を作成するには、関数 coder.newtype を使用します。

A = coder.newtype('double',[12,14 10],[0 0]);
B = coder.newtype('double',[14,16 10],[0 0]);
C = coder.newtype('double',[12,16 10],[0 0]);
alpha = 0.3;
beta = 0.6;
inputs = {A,B,C,alpha,beta};

CUDA ライブラリを生成するには、関数 codegen を使用します。

cfg = coder.gpuConfig('lib');
cfg.GpuConfig.EnableCUBLAS = true;
cfg.GpuConfig.EnableCUSOLVER = true;
cfg.GenerateReport = true;
codegen -config cfg-args inputs myStridedMatMulAdd

生成された CUDA コードには、入力および出力行列を初期化するためのカーネル myStridedMatMulAdd_kernelNN が含まれます。さらに、コードには cuBLAS ライブラリに対する cublasDgemmStridedBatched API 呼び出しも含まれます。以下のコードは、生成されたコードのスニペットです。

//
// File: myStridedMatMulAdd.cu
...

void myStridedMatMulAdd(const double A[1680], const double B[2240], const double
  C[1920], double alpha, double beta, double D[1920])
{
  double alpha1;

..alpha1 = alpha;
  beta1 = beta;
  cudaMemcpy(gpu_C, (void *)&C[0], 15360ULL, cudaMemcpyHostToDevice);
  myStridedMatMulAdd_kernel1<<<dim3(4U, 1U, 1U), dim3(512U, 1U, 1U)>>>(*gpu_C,
    *gpu_D);
  cudaMemcpy(gpu_alpha1, &alpha1, 8ULL, cudaMemcpyHostToDevice);
  cudaMemcpy(gpu_A, (void *)&A[0], 13440ULL, cudaMemcpyHostToDevice);
  cudaMemcpy(gpu_B, (void *)&B[0], 17920ULL, cudaMemcpyHostToDevice);
  cudaMemcpy(gpu_beta1, &beta1, 8ULL, cudaMemcpyHostToDevice);
  cublasDgemmStridedBatched(getCublasGlobalHandle(), CUBLAS_OP_N, CUBLAS_OP_N,
    12, 16, 14, (double *)gpu_alpha1, (double *)&(*gpu_A)[0], 12, 168, (double *)
    &(*gpu_B)[0], 14, 224, (double *)gpu_beta1, (double *)&(*gpu_D)[0], 12, 192,
    10);
  cudaMemcpy(&D[0], gpu_D, 15360ULL, cudaMemcpyDeviceToHost);
...
}

入力引数

すべて折りたたむ

オペランド。ベクトルまたは行列として指定します。A の列数は B の行数と等しくなければなりません。A の行数は C の行数と等しくなければなりません。B の列数は C の列数と等しくなければなりません。

データ型: double | single | int8 | int16 | int32 | int64 | uint8 | uint16 | uint32 | uint64
複素数のサポート: あり

名前と値の引数

オプションの引数のペアを Name1=Value1,...,NameN=ValueN として指定します。ここで、Name は引数名で、Value は対応する値です。名前と値の引数は他の引数の後に指定しなければなりませんが、ペアの順序は重要ではありません。

R2021a より前では、コンマを使用して名前と値をそれぞれ区切り、Name を引用符で囲みます。

例: D = gpucoder.stridedMatrixMultiplyAdd(A,B,C,'alpha',0.3,'beta',0.6,'transpose','CC');

A との乗算に使用されるスカラーの値。既定値は 1 です。

C との乗算に使用されるスカラーの値。既定値は 1 です。

2 つの文字からなる文字ベクトルまたは string で、行列乗算の前に行列 A および B で実行される演算を示します。可能な値は、標準 ('N')、転置 ('T')、または複素共役転置 ('C') です。

出力引数

すべて折りたたむ

積。スカラー、ベクトル、または行列として返されます。配列 D には、入力 A と同じ数の行と、入力 B と同じ数の列があります。

バージョン履歴

R2020a で導入

参考

アプリ

関数

オブジェクト

トピック