計算モード

計算モード#

BLAS レベル 3 ルーチンと拡張機能は代替計算モードをサポートしており、異なる数値プロパティーや精度の低下と引き換えにパフォーマンスを向上させることができます。

許可されたモードのリストは、コンパイル時、呼び出しごとまたはソースファイルごとに、または実行時に MKL_BLAS_COMPUTE_MODE 環境変数で指定できます。oneMKL は、ルーチン・パラメーターとハードウェア特性を考慮して、このリストから適切な実装を自動的に選択します。許可された代替モードのいずれも、選択されたデバイス上の指定されたルーチンでサポートされていない場合、または許可された代替モードのいずれもパフォーマンスの向上を期待できない場合、oneMKL は自動的に標準実装にフォールバックします。

注

特定のモードがパフォーマンス向上をもたらすかどうかは、ルーチンの種類、行列のサイズ、転置パラメーター、ハードウェア構成など、さまざまな要因によって異なります。現在の oneMKL リリースでは、実数単精度の場合は選択されたハードウェア上で gemm、gemmt、syrk、および syr2k の代替実装が利用可能であり、複素数単精度の場合は gemm のみが利用可能です。

デフォルトでは、oneMKL は代替計算モードを有効にしません。MKL_BLAS_COMPUTE_MODE 環境変数は、代替計算モードがアプリケーションにパフォーマンス上の利点と許容可能な精度をもたらすか迅速に評価することを目的としています。初期テストの後、呼び出しごとまたはソースファイルごとの API を使用して、代替モード設定をアプリケーション内で永続的に適用できます。

GPU で実行する場合、oneMKL の詳細出力には、各呼び出しにどのモードが使用されているか (標準モードか代替モードのいずれか) が示されます。詳細については、使用されているモードを確認を参照してください。

モードの設定#

利用可能な代替モードについては、以下の表で説明します。複数のモードを組み合わせることができるため、oneMKL は最高のパフォーマンスが得られると予想される許可されたモードのいずれかを選択できます。

呼び出しごとまたはソースファイルごとのモードの場合、次の表の 1 つ以上の oneapi::mkl::blas::compute_mode 値を OR 演算することによって、代替の計算モードが選択されます。 例:

using oneapi::mkl::blas; 
auto mode_settings = compute_mode::float_to_bf16x2 | compute_mode::float_to_tf32; /* これら 2 つのモードのいずれかを許可 */

MKL_BLAS_COMPUTE_MODE 環境変数を使用してグローバルな変更を行う場合、複数のモードはコンマで結合されます。

set MKL_BLAS_COMPUTE_MODE=FLOAT_TO_BF16X2,FLOAT_TO_TF32

compute_mode enum 値

環境変数の設定

説明

compute_mode::float_to_bf16

FLOAT_TO_BF16
単精度入力を内部的に bfloat16 形式に変換します。出力は単精度で累積されます。

精度は低下しますが、パフォーマンスは大幅に向上する可能性があります。

compute_mode::float_to_bf16x2

FLOAT_TO_BF16X2
各単精度入力値を内部的に 2 つの bfloat16 値の合計に変換します。出力は単精度で累積されます。

期待される精度は、標準の単精度演算と bfloat16 演算の精度の間になります。

注: 無限入力は出力に予期しない NaN を生成する可能性があります。

compute_mode::float_to_bf16x3

FLOAT_TO_BF16X3
各単精度入力値を内部的に 3 つの bfloat16 値の合計に変換します。出力は単精度で累積されます。

ほとんどの場合、期待される精度は標準の単精度演算と同等です。

注: 無限入力は出力に予期しない NaN を生成する可能性があります。

compute_mode::float_to_tf32

FLOAT_TO_TF32
それぞれの単精度入力値を内部的に tf32 形式に変換します。出力は単精度で累積されます。

期待される精度は float_to_bf16 と float_to_bf16x2 の間です。

compute_mode::complex_3m

COMPLEX_3M
標準的な複素乗算における 4 つの実数乗算を 3 つの実数乗算に減らします。

ほとんどの場合、期待される標準の算術演算と同等です。

compute_mode::any

ANY

任意の代替計算モードを許可します。

compute_mode::standard

STANDARD

任意の代替計算モードを許可しません。

compute_mode::prefer_alternate

PREFER_ALTERNATE
1 つ以上の代替モードと組み合わせて使用​​されます。

パフォーマンスが低下する可能性があっても、利用可能な場合は標準実装よりも代替計算モードを優先します。

compute_mode::force_alternate

FORCE_ALTERNATE
1 つ以上の代替モードと組み合わせて使用​​されます。

標準実装は絶対に使用しないでください。許可された代替実装が利用できない場合は、例外がスローされます。

呼び出しごとのモードの設定#

BLAS レベル 3 ルーチンと拡張機能は、パラメーター・リストの最後に、目的のモード設定を指定するオプションの oneapi::mkl::blas::compute_mode 引数をサポートします。USM API の場合、compute_mode 引数は、入力依存関係のリスト (存在する場合) の前に配置されます。どちらの引数も省略できます。以下に例を示します。

using oneapi::mkl::blas; 

sycl::buffer<float, 1> a_buffer, c_buffer; 
float *a_ptr, *c_ptr; 
/* ...*/ 

// バッファー API 
syrk(my_queue, n, k, uplo, trans, alpha, a_buffer, lda, beta, c_buffer, ldc, compute_mode::float_to_bf16); 

// バッファー API、float_to_bf16 モードを強制 
syrk(my_queue, n, k, uplo, trans, alpha, a_buffer, lda, beta, c_buffer, ldc, compute_mode::float_to_bf16 | compute_mode::force_alternate);

// 依存関係のない USM API 
syrk(my_queue, n, k, uplo, trans, alpha, a_ptr, lda, beta, c_ptr, ldc, compute_mode::float_to_bf16); 

// 依存関係がある USM API 
syrk(my_queue, n, k, uplo, trans, alpha, a_ptr, lda, beta, c_ptr, ldc, compute_mode::float_to_bf16, {event1, event2}); 

// USM API、依存関係はあるが特別な compute_mode 設定はない 
syrk(my_queue, n, k, uplo, trans, alpha, a_ptr, lda, beta, c_ptr, ldc, {event1, event2});

ソースファイルごとのモードの設定#

いずれかの MKL ヘッダーファイルをインクルードする前に MKL_BLAS_COMPUTE_MODE マクロを定義することで、ソースファイル内のすべての呼び出しに対してデフォルトのモードを設定できます。このマクロは、oneapi::mkl::blas::compute_mode タイプの式に設定する必要があります。

#define MKL_BLAS_COMPUTE_MODE oneapi::mkl::blas::compute_mode::complex_3m 
#include <oneapi/mkl.hpp> 

void my_function() { 
    /* ...*/ 

    // 3M モードはデフォルトで許可されます: 
    gemm(my_queue, m, n, k, trans_a, trans_b, alpha, a, lda, b, ldb, beta, c, ldc); 
}

oneMKL ルーチンに渡される compute_mode パラメーターは、デフォルト設定よりも優先されます。

#define MKL_BLAS_COMPUTE_MODE oneapi::mkl::blas::compute_mode::complex_3m 
#include <oneapi/mkl.hpp> 

void my_function() { 
    /* ...*/ 

    // 指定された compute_mode はデフォルトの 3M モードをオーバーライドします。 
    gemm(my_queue, m, n, k, trans_a, trans_b, alpha, a, lda, b, ldb, beta, c, ldc, compute_mode::standard); 
}

実行時のモードの設定#

MKL_BLAS_COMPUTE_MODE 環境変数を使用すると、実行時に代替計算モードを評価する方法として、アプリケーション全体のデフォルトの代替計算モードを設定できます。例: bash または同様のシェルの場合:

// my_application.cpp 
#include <oneapi/mkl.hpp> 

int main() { 
    /* ...*/ 

    // compute_mode 引数なしで gemm を呼び出す: 
    oneapi::mkl::blas::gemm(my_queue, /* ... */); 
}
export MKL_BLAS_COMPUTE_MODE=FLOAT_TO_BF16X3 
my_application # gemm は bf16x3 演算を使用する場合があります

呼び出しごとまたはソースファイルごとのモード設定は、MKL_BLAS_COMPUTE_MODE 環境変数よりも優先されます。この結果、-DMKL_BLAS_COMPUTE_MODE=oneapi::mkl::blas::compute_mode::standard を指定してアプリケーションをコンパイルすると、環境変数が無効になります。

使用されているモードを確認#

GPU では、oneMKL の詳細モードが有効になっている場合、各呼び出しで有効となり使用されている計算モードに関する情報が詳細ログに出力されます。たとえば、次の呼び出しでは、float_to_bf16 と float_to_tf32 の両方が有効になっており、float_to_bf16 が選択されています (分かりやすくするため一部の出力は省略されています)。

MKL_VERBOSE oneapi::mkl::blas::column_major::gemm[float](0x7ffd39046350,...,float_to_bf16|float_to_tf32) mode:float_to_bf16 host:nan device:nan GPU0

Verbose モードは、MKL_VERBOSE 環境変数を 1 に設定するか、mkl_verbose によって有効にできます。詳細については、インテル® ソフトウェア・ドキュメント・ライブラリー (英語) から入手できる インテル® oneAPI マス・カーネル・ライブラリー・デベロッパー・ガイドの「oneMKL Verbose モードを使用」を参照してください。