Intel oneAPIでC++ with SYCLプログラミングをはじめよう
はじめに
C++erの皆さん、SYCL使ってますか？そもそもSYCLを知っていますか？SYCLを使うとC++でGPGPUを実現できます。GPGPUといえばCUDAを連想するかもしれませんが、CUDAはNVIDIA専用なのに対し、SYCLはベンダーに依存しないという違いがあります。
昨今のGPGPUプログラミングではほとんどの場合NVIDIA製のGPUが使われており、AMDやIntelのGPUの名前が上がることはまずありません。これは多くのGPGPUソフトウェアがCUDAを使用しているからとも言えるでしょう。このせいでNVIDIAが市場を独占しており、適切な競争がなされない状態が続いています。
そんな中Pythonを見ると、AI分野において用いられるPyTorchやJAXなどのライブラリにおいてOpenXLAのおかげでNVIDIAに限らず様々なデバイスでの学習・推論が可能な状況が作られています。実際に使われているかはさておき、NVIDIA以外のGPUを選択肢に入れることができるような状況が出来上がったことは大変いいことです。
さらに、最近は主にゲーム業界を中心にAMDのRadeonが選択肢に上がることが増えているようです。RTX5000番代のドライバに不具合が多いことなどからNVIDIAを離れる選択を取るユーザーが出てきています。
つまり、NVIDIAに依存しないGPGPUを始めるのにとてもいい流れが来ています。これを機に、SYCLを使ってNVIDIAに依存しないGPGPUをはじめましょう！
この記事では、主にC++を使っている方向けに、Intel oneAPIを用いたSYCLの環境構築、サンプルコード、ビルド方法を簡単に説明します。SYCLは成長途中ということもあり、過去に書かれた記事では古い情報が多いので、2025年8月現在の情報をまとめました。
SYCLとは
SYCL(System-wide Compute Language)は、Vulkanなどで有名なKhronosグループが策定しているC++でヘテロジニアスプログラミングを行うための規格です。 先程はGPGPUのみを話題に出しましたが、実はこれにとどまりません。CPUやGPU、FPGAなどの様々なデバイスをC++から扱うことができます。CUDAでは、CUDAカーネルの記述はcuファイルに分ける必要がありましたが、SYCLでは同じcppファイルに記述することができます。C++の標準規格への統合を目指しているようです。
また、OpenCLでは、CUDAと比較して性能が低いという課題がありましたが、SYCLでは実装によってはCUDAと同程度の性能が出ることが知られています。
現在の主な実装として、Intel oneAPI DPC++やAdaptiveCppがあります。この記事では、oneAPIを使用する方法を説明します。
Intel oneAPIの環境構築
Dockerを使う場合
Dockerを使う場合は、Intelが以下のイメージを提供しています。 こちらをベースイメージとして用いるのが手っ取り早いと思います。VSCodeのDevContainerを使う場合などに便利です。
直接インストールする場合
直接インストールする場合は以下の記事を参照してください LinuxとWindowsそれぞれの説明がついているので、環境に合わせて従ってください。基本的にはIntel oneAPI Base Toolkitを選べば問題ありません。試していませんが、Intel C++ Essentialsでも大丈夫だと思います。
oneAPIは、インストールするだけではパスが通らないようになっています。これは、複数のバージョンが混ざらないようにしたり、Clangなどの既存のソフトウェアと干渉しないようにするためです。oneAPIを使用するときには、シェルでsetvarsスクリプトを走らせます。
Linuxの場合は以下を実行してください。
. /opt/intel/oneapi/setvars.sh
Windowsの場合は以下を実行してください
"C:\Program Files (x86)\Intel\oneAPI\setvars.bat"
Powershellを使いたい場合は以下になります。
cmd.exe "/K" '"C:\Program Files (x86)\Intel\oneAPI\setvars.bat" && powershell'
詳しくは以下を参照してください。
GPU環境の構築
oneAPIはデフォルトではOpenCLのバックエンドのみ使えます。しかし、これはパフォーマンスが低いので、NVIDIAやAMD向けにバックエンドが別に提供されています。
NVIDIAを使用する場合はoneAPI for NVIDIA GPUsをインストールします。
こちらからインストールスクリプトをダウンロードして実行してください。
事前にCUDAをインストールしておく必要があります。
詳しくは以下を参照してください。
AMDでは同様にoneAPI for AMD GPUsをインストールします。
こちらからインストールスクリプトをダウンロードして実行してください。
事前にROCmをインストールしておく必要があります。
詳しくは以下を参照してください。
SYCLで使用可能なデバイスの一覧はsycl-lsコマンドで調べられます。以下のように出力されるので、使用したいバックエンドとデバイスが認識されているか確認してください。
# sycl-ls
[cuda:gpu][cuda:0] NVIDIA CUDA BACKEND, NVIDIA GeForce RTX 2060 SUPER 7.5 [CUDA 12.6]
[opencl:cpu][opencl:0] Intel(R) OpenCL, Intel(R) Core(TM) i7-9700 CPU @ 3.00GHz OpenCL 3.0 (Build 0) [2025.20.6.0.04_224945]
サンプルコード
SYCLを用いたGPGPUのサンプルコードです。
ものすごく簡単に言うと、コマンドキューを作成し、これにカーネルを送信するとGPU上で実行されます。データのやり取りはいい感じのバッファークラスを使って行います。
#include <iostream>
#include <sycl/sycl.hpp>
#include <vector>
int main() {
const int N = 10;
// コマンドキューを作成
sycl::queue queue(sycl::default_selector_v);
// コマンドキューのデバイス情報を表示
std::cout << "Device: "
<< queue.get_device().get_info<sycl::info::device::name>()
<< std::endl;
std::vector<int> vec(N);
// ホストとデバイスでデータをやり取りしてくれるSYCL Bufferを定義
sycl::buffer<int, 1> buf(vec.data(), vec.size());
// カーネルを送信
queue.submit([&](sycl::handler &cgh) {
// Accessorを定義してデバイスで領域を確保
sycl::accessor acc(buf, cgh, sycl::write_only);
cgh.parallel_for(sycl::range<1>(vec.size()), [=](sycl::item<1> item) {
const auto id = item.get_linear_id();
acc[id] = id * id;
});
});
// 明示的にホスト側にデータをコピー
buf.get_host_access();
for (const auto v : vec) {
std::cout << v << ", ";
}
std::cout << std::endl;
return 0;
}
出力は以下のようになります。ビルド方法は次の章に書きます。
Device: NVIDIA GeForce RTX 2060 SUPER
0, 1, 4, 9, 16, 25, 36, 49, 64, 81,
SYCLプログラミングの詳しい解説は、以下のSYCL Academyがおすすめです(スライドがhtmlで書かれているので、英語の苦手な方もGoogle翻訳などで読めます)。
ビルド
詳しい説明はこちらを参照してください。
コマンドラインによるビルド
oneAPIではIntel oneAPI Compilerを使ってコンパイルします。
単にC++のソースコードをコンパイルする場合、g++やclang++と同じようにicpxコマンドでコンパイルできます。
icpx main.cpp
SYCLを使う場合は-fsyclオプションをつけます
icpx -fsycl main.cpp
CPU以外のデバイスを使用する場合は、使用したいデバイスのアーキテクチャを指定します。
デフォルトではspir64が使用されます。例えば、NVIDIAのCUDAバックエンドを使いたい場合は以下のようになります。
icpx -fsycl -fsycl-targets=spir64,nvptx64-nvidia-cuda main.cpp
詳しくは以下を参照してください。
CMakeによるビルド
一定以上の規模のソフトウェアを作る場合はCMakeを使うことになるかと思います。
CMAKE_C_COMPILERにicxを、CMAKE_CXX_COMPILERにicpxを指定します。
find_packageでIntelSYCLパッケージを読み込みます(古い記事だとIntelDPCPPとなっていますが、これはDeprecatedになりました)。
SYCLのカーネルが含まれるcppファイルを、add_sycl_to_targetで指定します。
target_compile_optionsとtarget_link_optionsで-fsycl-targetsを指定します。
以下サンプルです。
set(CMAKE_CXX_COMPILER icpx)
set(CMAKE_C_COMPILER icx)
cmake_minimum_required(VERSION 3.25)
project(main LANGUAGES CXX)
find_package(IntelSYCL REQUIRED)
add_executable(main main.cpp)
add_sycl_to_target(TARGET main SOURCES main.cpp)
target_compile_options(main PRIVATE
-fsycl-targets=spir64,nvptx64-nvidia-cuda)
target_link_options(main PRIVATE
-fsycl-targets=spir64,nvptx64-nvidia-cuda)
詳しくは以下を参照してください。
最後に
CPUではAMDがIntel一強の時代を破りました。GPUにおいては長らくNVIDIA一強の時代が続き、まだしばらくは続きそうですが破られる可能性はすでに十分にあると思います。
C言語やC++は、アセンブリとは異なりCPUのアーキテクチャに依存しないプログラミングを実現しました。アセンブリを頑張って書いたほうが極限の性能を出せるはずですが、今日日アセンブリを手書きする機会が少ないのは、それよりもアーキテクチャ非依存性や言語の高級さがもたらす利益が大きいからだと思います。
GPUにおいても同じことが言えるのではないでしょうか。CUDAやROCmなどのベンダーに依存したフレームワークで直接書けば、そのデバイスで極限の性能をひきだすことはできるでしょう。しかし、それ以上にベンダーの選択肢を広げられることや、(今回は強調しませんでしたが)より高次元なプログラミングパラダイムを使えることは、より大きな利益をもたらすのではないかと思います。
しかしSYCLの認知度が低いままでは、やがてSYCLも廃れてしまいます。私としてはそれは非常にもったいないと思います。CUDAで事足りてしまう現状ではなかなか腰を上げにくいと思いますが、SYCLを皆さんの技術選定の選択肢に入れていただければありがたい限りです。
Discussion