インテル® DPC++ 互換性ツール
CUDA* コードから DPC++ への移行

内容•

oneAPI の概要•

CUDA* と SYCL*•

インテル® DPC++ 互換性ツールの移行の流れ•

ベクトル加算の移行例•

実際の移行例 (最も一般的な手法)•Eclipse* と Visual Studio* への統合

•

まとめ•

デモ・チュートリアル•

簡単な CUDA* ファイル・プロジェクト•

複数の CUDA* ファイル・プロジェクトを移行

2

oneAPI 業界
イニシアチブアプリケーションのワークロードには多様なハードウェアが必要

独自な開発環境から脱却

ミドルウェアとフレームワーク...
C++ と SYCL* 標準をベースとするのクロスアーキテクチャー言語

コミュニティーと業界の協力を促進するオープンな活動
すべてのアーキテクチャーおよびベンダー間でコードを再利用

oneAPI 業界仕様
ダイレクト・プログラミング

API ベースのプログラミング
ライブラリー

データ並列 C++

ドメイン固有の関数を高速化するために設計された強力なライブラリー

算術

スレッド化

DPC++ ライブラリー

解析 / MLDNN

ML 通信ビデオ処理

低レベルのハードウェア・インターフェイス

低レベルなハードウェア
抽象化レイヤー
CPU

GPU

FPGA

ほかのアクセラレーター

独自プログラミング・モデルの経済的および技術的負担から、ヘテロジニアス・コンピューティングを解放する生産的でスマートな方法を提供します3インテル® oneAPI 製品
インテルの豊富な CPU ツール資産をXPU に拡張

アプリケーションのワークロードには多様なハードウェアが必要

ミドルウェアとフレームワーク...
高度なコンパイラー、ライブラリー、および移植、解析、デバッグツールを含む完全なセット

インテル® oneAPI 製品▪ 先端のハードウェア機能を活用してコンピューティングを加速▪ 従来のプログラミング・モデルおよびコードベース(C++、Fortran、Python*、OpenMP* など) と相互運用可能であるため、開発者は既存のアプリケーションと oneAPI のシームレスな連携を確信できます

互換性ツール▪ 単一のコードベースにより新しいシステムやアクセラレーターへの移行を容易にし、開発者はより多くの時間をイノベーションに費やすことができます

言語

ライブラリー

解析とデバッグツール

低レベルのハードウェア・インターフェイス

XPU

CPUGPU

FPGA

こちらから入手できます (英語)
詳細については、software.intel.com/oneapi (英語) をご覧ください。
一部の機能はアーキテクチャーごとに異なるため、カスタム・チューニングが必要です。ほかのアクセラレーターも今後サポートされる可能性があります。

4

5

インテル® DPC++ 互換性ツール (インテル® DPCT)
コードの移行時間を最小化

すでに CUDA* で記述されているコードを DPC++ に移行する開発者を支援、可能な場合は解読可能なコードを生成しますインテル® DPC++ 互換性ツールの作業の流れ

80-90% を変換

通常はコードの 80-90% を自動的に移行します
開発者がアプリケーションの移植を完了できるように支援するインラインコメントを提供します

コーディングを完了して目的のパフォーマンスにチューニング

インラインコメント付きの人間が解読可能な
DPC++ コード

開発者の CUDA*
ソースコード互換性ツールDPC++ ソースコード

5インテル® oneAPI
ベース・ツールキット
データ中心のワークロードを加速
インテル® CPU、GPU、FPGA 上で高性能アプリケーションを開発するコアツールとライブラリーのセット

対象▪
▪

ダイレクト・プログラミング

API ベースのプログラミング

解析/デバッグツール

インテル® oneAPI
DPC++/C++ コンパイラー

インテル® oneAPI
DPC++ ライブラリー(インテル® oneDPL)

インテル® VTune™
プロファイラーインテル® DPC++
互換性ツールインテル® oneAPI
マス・カーネル・ライブラリー(インテル® oneMKL)

インテル® Advisor

インテル® ディストリビューションの Python*

インテル® oneAPI データ・アナリティクス・ライブラリー(インテル® oneDAL)

インテル® ディストリビューションの GDB

oneAPI ベース・ツールキット用
インテル® FPGA アドオン

インテル® oneAPI スレッディング・ビルディング・ブロック(インテル® oneTBB)

さまざまさ業界の幅広い開発者
アドオン・ツールキットを使用する開発者

主な機能/利点▪

インテル® oneAPI ベース・ツールキット

データ並列 C++ コンパイラー、ライブラリーおよび開発ツール▪

CUDA* で記述された既存のコードの移行を支援する
DPC++ 互換性ツール▪

Python* ディストリビューションには、高速化された
scikit-learn*、NumPy*、SciPy* ライブラリーが含まれます▪

スレッド化、数学、データ解析、ディープラーニング、ビデオ/画像/信号処理向けに最適化されたパフォーマンス・ライブラリー

詳細: intel.com/oneAPI-HPCKit (英語)インテル® oneAPI ビデオ・プロセシング・ライブラリー(インテル® oneVPL)
インテル® oneAPI コレクティブ・コミュニケーション・ライブラリー(インテル® oneCCL)
インテル® oneAPI ディープ・
ニューラル・ネットワーク・ライブラリー (インテル® oneDNN)
インテル® インテグレーテッド・
パフォーマンス・プリミティブ(インテル® IPP)

6

CUDA* と SYCL*

7

なぜ SYCL* へ移行するのか• コードの移植性- Nvidia*、インテル、AMD* のターゲット GPU- CPU、FPGA、多様なハードウェア• 最小限のオーバーヘッド• 皆さんは CUDA* の使い方はすでに知っています8

なぜ SYCL* へ移行するのか• コードの移植性- Nvidia*、インテル、AMD* のターゲット GPU- CPU、FPGA、多様なハードウェア• 最小限のオーバーヘッド• 皆さんは CUDA* の使い方はすでに知っています• それは基本的には SYCL* の使い方を理解しているとも言えます9

CUDA* から SYCL* への移行は容易• 大部分の CUDA* コンセプトは、1 対 1 で SYCL* コンセプトに対応します• SYCL* 2020 仕様に詳細が示されています• 同様のパフォーマンス最適化を適用できます• 同様のプロファイル・ツールが利用できます (nvprof、nsys、ncu など)

10

考えを切り替えましょう• SYCL* には GPU の伝統と資産があります• ほとんどの CUDA* コンセプトは SYCL* と等価です• SYCL* は CUDA* や他のコンセプトの上位にある抽象化レイヤです11しかしまったく同じではありません• SYCL* はアウトオブオーダー・キューを提供します• SYCL* ではバッファーも利用できます• SYCL* はタスクグラフに注目します• 問題の合計サイズはグリッドとブロックで表現しません- オプションで work-group (ブロック) サイズを指定します12

問題サイズの表現の違い• SYCL* ローカルレンジ = ブロックごとの CUDA* スレッド
• SYCL* グローバルレンジ =グリッドごとの CUDA* ブロック x ブロックごとのスレッド• ローカルレンジの選択を SYCL* ランタイムに任せることができます
CUDA*

SYCL*

ハードウェアへのマッピング

スレッドブロック

ワークグループ
sycl::group

同じ「場所」で実行されるハードウェア・スレッドのチーム

ワープ

サブグループ
sycl::sub_group

ハードウェア・スレッド

スレッド

ワーク項目
sycl::nd_item

SIMD レーン (または、ハードウェア・スレッドごとにインターリーブされた「実行スレッド」)
13

CUDA* は列優先 vs SYCL* は行優先
4 つのワーク項目のサブグループ

ワークグループの次元 Y

CUDA* の Y 次元最大
レンジは、 X と Z よりも大きい

ワークグループ
ワークグループ の次元 Zの次元 X

SYCL* の Z 次元最大レンジは、X と Y よりも大きい

ワークグループ
14

ベクトル加算の移行例

15簡単な例を移行▪ dpct [オプション] [<ソース 0>...<ソース N>]▪ サポートされる CUDA* ヘッダーファイルが利用可能であること▪ -cuda-include-path オプションで指定できます▪ ツールのヘルプ▪ dpct --help

16

ベクトル加算の例: インテル® DPCT を使用して移行
CUDA

DPC++
1
2

3
4
https://github.com/oneapi-src/oneAPI-samples/tree/master/Tools/Migration/vector-add-dpct (英語)17ベクトル加算の例 (続き)
CUDA

DPC++

5

6
7
8
https://github.com/oneapi-src/oneAPI-samples/tree/master/Tools/Migration/vector-add-dpct (英語)18

実際の移行例 (最も一般的な手法)

19

移行の流れ簡単なプロジェクトから複雑なプロジェクトにおける典型的な準備ステップ

準備

インターセプト・ビルド

移行dpct確認

確認して手動で編集20

20

インターセプト・ビルド▪ インターセプト・ビルドによりコンパイル・データベースを作成します▪ Make や Cmake を使用するプロジェクト向け▪ コンパイルオプション、設定、マクロ定義、インクルード・ディレクトリーなどを追加します▪ ビルドコマンドを含む JSON ファイルを作成します▪ “intercept-build” の前に “make clean” を実行します21インテル® DPCT の基本オプション▪ dpct [オプション] [<ソース 0>...<ソース N>]
インテル® DPCT の基本オプション--in-root

移行するソースファイルを含むディレクトリーのルートパス--out-root

生成するファイルを格納するディレクトリーのルートパス-p

データベースをコンパイルする JSON ファイルへのパス--process-all

マイグレーター/コピーは、--in-root ディレクトリーから --out-root ディ
レクトリーにすべてのファイルをコピーするため、.cu ファイルを個別に指定する必要はありません--extra-arg

追加の Clang コンパイラー・オプションを指定します
例: dpct --extra-arg=”-std=c++14” –extra-arg=”-I…”--format-style

出力ファイルのフォーマット形式を指定します
例: =llvm、=google、=custom (.clang-format ファイルを使用)--format-range

コードがない (=none)、移行されたコード (=migrated)、またはすべてのコー
ド (=all) に適用されるコード・フォーマット
22

インテル® DPCT の推奨オプション▪ dpct [オプション] [<ソース 0>...<ソース N>]
移行とデバッグを容易にするインテル® DPCT オプション--keep-original-code

生成される DPC++ ファイルにオリジナルの CUDA* コードを残しますオリジナルの CUDA* コードと生成された DPC++ コードを簡単に比較できます--comments

生成されたコードの説明をコメントとして挿入します---always-use-asynchandler

非同期の例外ハンドラーで常に cl::sycl::queue を生成します▪ ほかにもオプションはあります。dpct --help コマンドで確認してください23dpct 名前空間の使い方▪ dpct 名前空間は、入力ソースコードの移行を支援するヘルパー関数やマクロを提供します▪ dpct::▪ ヘッダーファイルで実装されます (include/dpct)▪ 生成されるコードの一部になります▪ 例: dpct_malloc、dpct_memcpy、get_buffer、get_default_queue、get_default_context▪ 新規に DPC++ コードを記述する場合の使用は推奨されません

24最も一般的な手法 (BKM)• 漸進的に移行します•

長い CUDA* ソースを一度に移行する際に dpct が複数のエラーを報告する
場合、1 つずつ実行してください• “intercept-build make” を実行する前に “make clean” でプロジェクトをクリーンアップします• intercept-build make -k を実行すると、コンパイル・データベースを作成する際にターゲットを生成できない場合でも処理を続行できますインテル® ソフトウェア製品のパフォーマンス/最適化に関する詳細は、最適化に関する-注意事項を参照してください。
25

移行前のコード変更▪ ソースファイルに構文の誤りがないことを確認します▪ clang と nvcc の違いにより修正が必要になる場合があります
1. clang パーサーを使用する特定のシナリオでは、名前空間の変更が必要になることがあります
2. clang では追加のフォワードクラス宣言が必要になる場合があります

3. カーネルのトリプルブラケット内のスペースは nvcc では許容されますがclang では許されません。
▪

▪

例: cuda_kernel<< <num_blocks, threads_per_block>> >(args…)

詳細については、llvm.org の clang を使用した CUDA* のコンパイル (英語) を参照してください26

統合共有メモリー (USM) の使い方
▪ DPC++ は、ポインターベースのアプローチでホストとデバイスのメモリーを管理できる USM をサポートします▪ USM では SYCL* バッファーと比べて関連コードが少なくなります▪ 互換性ツールはデフォルトで USM を使用します▪ インテル以外のハードウェアを対象とする他社のコンパイラーでは問題が生じる可能性があります
DPCT USM オプション--usm-level

統合共有メモリー (USM) のレベルを設定します=Restricted: USM を使用 (デフォルト)=none: ヘルパー関数と SYCL* バッファーを使用

27

診断リファレンス
互換性ツールは移行に関する問題をコードのコメントとして通知します/path/to/file:20:1: warning:
DPCT10XX:0: text of thewarning//source code line for which
warning was generated

互換性ツールの診断リファレンスをご覧ください (日本語)28コードの確認と修正が必要です
診断リファレンス (日本語)•

エラーのあるロジックが (*,0) コードに置き換えられたかコメントアウトされました•

等価な DPC++ API がありません•

CUDA* の計算機能に依存するロジックです•ハードウェア依存の API (clock()) です•

いつくかの API は移行ツールではサポートされません•

実行時間測定ロジックです•

組込みベクトル型の競合を処理しています•

cuBLAS API を移行しようとしています (引数リストを確認してください)インテル® ソフトウェア製品のパフォーマンス/最適化に関する詳細は、最適化に関する-注意事項を参照してください。
29

SPECFEM3D_GLOBE の例▪ SPECFEM3D_GLOBE は、全域および地域 (大陸規模) の地震波伝搬をシミュレートします▪ 公式リポジトリー:
https://github.com/geodynamics/specfem3d_globe (英語)

出典: specfem3d_globe-manual.pdf (英語)30SPECFEM3D_GLOBE – DPC++ への移行

準備移行

$ git clone --recursive --branch devel https://github.com/geodynamics/specfem3d_globe.git$ ./configure --with-cuda=cuda9 CUDA_LIB=${CUDA_ROOT}/lib64/ ¥CUDA_INC=${CUDA_ROOT}/include/ MPI_INC=${I_MPI_ROOT}/include/$ intercept-build make -i$ dpct -p compile_commands.json

リファレンス (日本語) を使用して診断メッセージを確認して手動で編集します

確認

ほかの明確ではない問題に対処します31

診断メッセージの例
DPCT {診断番号} (count)

まとめ

DPCT1000 (6)、DPCT1001(6)、DPCT1003 (111)、 DPCT1009 (8)、DPCT1010 (3)、 DPCT1024 (2)エラー処理の各種シナリオ

DPCT1005 (4)、 DPCT1012 (4)、
DPCT1017 (10)、 DPCT1019 (1)、
DPCT1022 (1)、 DPCT1026 (5)、DPCT1027 (3)、 DPCT1051 (4)SYCL* で等価な API を利用できません(例: デバイスのバージョン、特定のデバイス・プロパティー、タイミングロジックなど)

DPCT1039 (9)アトミック処理 (デフォルトではグローバルアトミック、ローカルでは介入が必要)

DPCT1049 (59)使用するワークグループ・サイズを検証してく
ださい
32

IDE との併用

33

Eclipse*: ガウシアンの例

インテル® ソフトウェア製品のパフォーマンス/最適化に関する詳細は、最適化に関する-注意事項を参照してください。

34

Visual Studio* 2019: ガウシアンの例

インテル® ソフトウェア製品のパフォーマンス/最適化に関する詳細は、最適化に関する-注意事項を参照してください。

35

まとめ
• oneAPI は統一されたプログラミング・モデルを提供し、多様なアーキテクチャーでの開発を簡素化します• インテル® DPC++ 互換性ツールは、CUDA* で記述されたコードをDPC++ 環境に移行する際に開発者を支援する、生産性向上に役立つツールです• DPC++ は、並列処理を表現するポータブルでアーキテクチャーに依存しない、業界標準に基づいたオープン仕様の言語です

36

関連資料
▪ インテル® DPC++ 互換性ツール Jupyter チュートリアル (英語)▪ インテル® DPC++ 互換性ツール (英語)▪ ユーザーガイド (日本語)▪ 導入ガイド (日本語)▪ リリースノート (英語)37

チュートリアルによるNeedleman Wunsch とHydroC の移行例

38

Jupyter* チュートリアルの入手方法
https://software.intel.com/content/www/us/en/develop/articles/intel-dpcpp-compatibility-tool-training.html (英語)

ここからダウンロード

39Jupyter* チュートリアルの実行例

前のスライドで入手したチュートリアルを、ローカル Linux* システムで実行できます40

例: 簡単な CUDA* プロジェクトの移行• Rodinia Benchmark Suite v3.1 の紹介• インテル® DPC++ 互換性ツールの環境設定と確認
• デモ•

移行の計画

•

互換性ツールのオプション•

Needleman Wunsch アプリケーションの移行

http://rodinia.cs.virginia.edu/doku.php (英語)https://software.intel.com/en-us/get-started-with-intel-dpcpp-compatibility-tool (英語)https://software.intel.com/content/www/us/en/develop/documentation/intel-dpcpp-compatibility-tool-user-guide/top/migrate-a-project/migrate-a-project-on-linux.html (英語)https://software.intel.com/en-us/intel-dpcpp-compatibility-tool-user-guide-usage-workflow-overview (英語)41

例: HydroC - 複数の CUDA* ファイル・プロジェクトの移行• インテル® DPC++ 互換性ツールの環境設定と確認• デモ•

移行の計画: アプリケーションのファイル構成を理解します ...

•インターセプト・ビルドのオプション•

互換性ツールのオプション•

HydroC アプリケーションの移行

https://github.com/HydroBench/Hydro/tree/master/HydroC/cuHydroC_2DMpi/Src (英語)https://github.com/HydroBench/Hydro (英語)https://github.com/HydroBench/Hydro/blob/master/License.txt (英語)42oneAPI を試してみませんか?
1. テストの対象となるワークロードや候補を特定します
a. インテル® DPCT をダウンロードして、オンプレミスの DPC++ にコードを移行します
b. インテル® DevCloud は、最新のインテル® ハードウェアとインテル® oneAPI
ソフトウェアを利用できるクラウドベースの開発サンドボックスであり、コードやサンプルのテスト、チューニング、最適化を行うことができます

2. 詳細は、http://software.intel.com/oneapi (英語) をご覧ください。
ドキュメント、ダウンロード、インテル® DevCloud およびフォーラムへのアクセスが可能です43法務上の注意書きと最適化に関する注意事項
本資料には、開発中の製品、サービスおよびプロセスについての情報が含まれています。本資料に含まれる情報は予告なく変更されることがあります。 インテル® テクノロ
ジーの機能と利点はシステム構成によって異なり、対応するハードウェアやソフトウェア、またはサービスの有効化が必要となる場合があります。詳細については、各システ
ムメーカーまたは販売店にお問い合わせいただくか、http://www.intel.co.jp/ を参照してください。
上記のベンチマーク結果は、追加のテストによって変更が必要になる可能性があります。結果は、テストに使用される特定のプラットフォーム構成や作業負荷に依存します。
個々のユーザーのコンポーネント、コンピューター・システム、作業負荷では同様の結果が得られない可能性があります。結果は、必ずしもほかのベンチマークを代表するものではなく、ほかのベンチマークでは、結果が異なることがあります。
性能に関するテストに使用されるソフトウェアとワークロードは、性能がインテル® マイクロプロセッサー用に最適化されていることがあります。SYSmark* や
MobileMark* などの性能テストは、特定のコンピューター・システム、コンポーネント、ソフトウェア、操作、機能に基づいて行ったものです。結果はこれらの要因によって異なります。製品の購入を検討される場合は、他の製品と組み合わせた場合の本製品の性能など、ほかの情報や性能テストも参考にして、パフォーマンスを総合的に評価することをお勧めします。詳細については、www.intel.com/benchmarks (英語) を参照してください。
本資料に掲載されている情報は現状のまま提供され、いかなる保証もいたしません。本資料は、明示されているか否かにかかわらず、また禁反言によるとよらずにかかわらず、いかなる知的財産権のライセンスを許諾するためのものではありません。製品に付属の売買契約書『Intel's Terms and Conditions of Sale』に規定されている場合を除き、インテルはいかなる責任を負うものではなく、またインテル製品の販売や使用に関する明示または黙示の保証（特定目的への適合性、商品性に関する保証、第三者の特許権、著作権、その他、知的財産権の侵害への保証を含む）をするものではありません。
© 2021 Intel Corporation. 無断での引用、転載を禁じます。Intel、インテル、Intel ロゴ、その他のインテルの名称やロゴは、Intel Corporation またはその子会社の商標です。
Khronos* は登録商法であり、SYCL* は Khronos Group, Inc の商標です。
* その他の社名、製品名などは、一般に各社の表示、商標または登録商標です。

最適化に関する注意事項
インテル® コンパイラーでは、インテル® マイクロプロセッサーに限定されない最適化に関して、他社製マイクロプロセッサー用に同等の最適化を行えないことがあります。これには、インテル® ストリーミング SIMD 拡張命令 2、インテル® ストリーミング SIMD 拡張命令 3、インテル® ストリーミング SIMD 拡張命令 3 補足命令などの最適化が該当します。インテルは、他社製マイクロプロセッサーに関して、いかなる最適化の利用、機能、または効果も保証いたしません。本製品のマイクロプロセッサー依存の最適化は、インテル® マイクロプロセッサーでの使用を前提としています。インテル® マイクロアーキテクチャーに限定されない最適化のなかにも、インテル® マイクロプロセッサー用のものがあります。この注意事項で言及した命令セットの詳細については、該当する製品のユーザー・リファレンス・ガイドを参照してください。
改訂 #20110804

44