オープンCL

OpenCL API

原作者	アップル社
開発者	クロノスグループ
初回リリース	2009年8月28日; 15年前 (2009-08-28)
安定版リリース	3.0.17 [1]  / 2024年10月24日; 27日前 (24 October 2024)
書かれた	C++バインディング付きC
オペレーティング·システム	Android（ベンダー依存）、[2] FreeBSD、[3] Linux、macOS（Pocl経由）、Windows
プラットフォーム	ARMv7、ARMv8、[4] Cell、IA-32、Power、x86-64
タイプ	異機種コンピューティング API
ライセンス	OpenCL仕様ライセンス
Webサイト	クロノス

OpenCL ( Open Computing Language ) は、中央処理装置(CPU)、グラフィックス処理装置(GPU)、デジタル信号プロセッサ(DSP)、フィールド プログラマブル ゲート アレイ(FPGA)、その他のプロセッサまたはハードウェア アクセラレータで構成される異機種プラットフォーム間で実行されるプログラムを作成するためのフレームワークです。OpenCL は、これらのデバイスをプログラミングするためのプログラミング言語( C99ベース) と、プラットフォームを制御してコンピューティング デバイス上でプログラムを実行するためのアプリケーション プログラミング インターフェイス(API) を指定します。OpenCL は、タスクおよびデータ ベースの並列処理を使用した並列コンピューティングの標準インターフェイスを提供します。

OpenCLは、非営利のオープンスタンダード組織であるKhronos Groupによって維持されているオープンスタンダードです。準拠した実装（適合性テストスイートに合格したもの）は、AMD、ARM、Cadence、Google、Imagination、Intel、Nvidia、Qualcomm、Samsung、SPI 、Verisiliconなどのさまざまな企業から入手できます。^{[8] [9]}

OpenCL では、コンピューティング システムは、ホストプロセッサ (CPU)に接続された中央処理装置(CPU) またはグラフィックス処理装置 (GPU) などの「アクセラレータ」である多数のコンピューティング デバイスで構成されているとみなされます。プログラムを作成するための C のような言語が定義されています。OpenCL デバイスで実行される関数は「カーネル」と呼ばれます。^{[10] : 17}単一のコンピューティング デバイスは通常、複数の処理要素(PE)を含む複数のコンピューティング ユニットで構成されます。単一のカーネル実行は、すべての PE または多数の PE で並列に実行できます。コンピューティング デバイスをコンピューティング ユニットと PE に分割する方法は、ベンダー次第です。計算ユニットは「コア」と考えることができるが、コアの概念はOpenCLでサポートされているすべての種類のデバイス（または「CPU」のカテゴリ内でさえ）にわたって定義するのが難しい。^{[11] : 49–50}また、計算ユニットの数はベンダーのマーケティング資料に記載されているコアの数と一致しない可能性がある（実際にはSIMDレーンを数えている可能性がある）。^[12]

OpenCLは、C言語に似たプログラミング言語に加えて、アプリケーションプログラミングインターフェイス（API）を定義しています。これにより、ホストで実行されているプログラムがコンピューティングデバイスでカーネルを起動し、デバイスメモリを管理できるようになります。デバイスメモリは（少なくとも概念的には）ホストメモリとは分離されています。OpenCL言語のプログラムは実行時にコンパイルされることが意図されているため、OpenCLを使用するアプリケーションは、さまざまなホストデバイスの実装間で移植可能です。^[13] OpenCL標準は、 CおよびC++のホストAPIを定義しています。Python 、^[14] Java、Perl、[ ^15] D ^[16] .NET [ ^11]などの他のプログラミング言語やプラットフォーム用のサードパーティAPIが存在します。¹⁵ OpenCL標準の実装は、 CおよびC++のAPIを実装するライブラリと、対象となるコンピューティングデバイス用のOpenCL Cコンパイラで構成されます。

OpenCLプログラミングモデルを他の言語に公開したり、カーネルソースを検査から保護したりするために、標準ポータブル中間表現（SPIR）^[17]を、フロントエンドコンパイラとOpenCLバックエンドの間でカーネルを転送するためのターゲットに依存しない方法として使用できます。

最近では、クロノスグループは、プログラミングの生産性を向上させるために、純粋なC++17に基づく単一ソースのeDSLとしてOpenCLの高レベルプログラミングモデルであるSYCL ^[18]を承認しました。C++カーネルに興味があるがSYCLの単一ソースプログラミングスタイルには興味がない人は、「C++ for OpenCL」言語で書かれたコンピューティングカーネルソースでC++機能を使用することができます。^[19]

OpenCLは計算デバイスのための4レベルのメモリ階層を定義している: ^[13]

• グローバル メモリ: すべての処理要素によって共有されますが、アクセス レイテンシが高くなります ( __global )。
• 読み取り専用メモリ: 小さく、レイテンシが低く、ホスト CPU からは書き込み可能だが、コンピューティング デバイスからは書き込み不可 ( __constant )。
• ローカルメモリ: 処理要素のグループによって共有されます ( __local )。
• 要素ごとのプライベートメモリ (レジスタ; __private )。

すべてのデバイスがこの階層の各レベルをハードウェアで実装する必要はありません。階層内のさまざまなレベル間の一貫性は緩和されており、明示的な同期構造、特にバリアによってのみ強制されます。

デバイスはホストCPUとメモリを共有する場合と、共有しない場合があります。^[13]ホストAPIは、デバイスのメモリバッファの ハンドルと、ホストとデバイス間でデータを転送する機能を提供します。

計算カーネルの記述に使用されるプログラミング言語は、カーネル言語と呼ばれます。OpenCL は、デバイス上で実行されるカーネル計算を指定するためにC / C++ベースの言語を採用していますが、アクセラレータの異種ハードウェア リソースへの効率的なマッピングを容易にするために、いくつかの制限と追加事項があります。従来、OpenCL 標準でアクセラレータをプログラムするために OpenCL C が使用されていましたが、その後、OpenCL C のすべての機能を継承しながらもカーネル ソースで C++ 機能を使用できるようにした OpenCL カーネル言語用の C++ が開発されました。

OpenCL C ^[20]は、OpenCLのデバイスモデルに適合するように適合されたC99ベースの言語方言です。メモリバッファはメモリ階層の特定のレベルに存在し、ポインタにはこれを反映して領域修飾子__global、__local、__constant、および__privateが付けられます。デバイスプログラムがメイン関数を持つ代わりに、OpenCL C関数は__kernelとマークされ、ホストプログラムから呼び出されるプログラムへのエントリポイントであることを示します。関数ポインタ、ビットフィールド、および可変長配列は省略され、再帰は禁止されています。^[21] C標準ライブラリは、数学プログラミング向けのカスタム標準関数セットに置き換えられました。

OpenCL C は、ベクトル型と演算、同期、および作業項目と作業グループを操作する関数による並列処理の使用を容易にするように拡張されています。 ^[21]特に、C の対応する型と同様に動作するfloatやdouble などのスカラー型に加えて、OpenCL はfloat4 (単精度浮動小数点の 4 次元ベクトル) などの固定長ベクトル型を提供します。このようなベクトル型は、さまざまな基本型に対して 2、3、4、8、16 の長さで使用できます。^{[20] : § 6.1.2}これらの型のベクトル化された演算は、CPU 上で OpenCL プログラムを実行するときに、 SSEやVMXなどのSIMD命令セットにマップすることを目的としています。^[13]その他の特殊な型には、2 次元および 3 次元の画像型があります。^{[20] : 10–11}

以下は、OpenCL C の 行列とベクトルの乗算アルゴリズムです。

// A*x を乗算し、結果を y に残します。// 
A は行優先の行列で、(i,j) 要素は A[i*ncols+j] にあります。
__kernel void matvec ( __global const float * A , __global const float * x , uint ncols , __global float * y ) { size_t i = get_global_id ( 0 ); // 行インデックスとして使用されるグローバル ID __global float const * a = & A [ i * ncols ]; // i 番目の行へのポインターfloat sum = 0.f ; // ドット積の累算器for ( size_t j = 0 ; j < ncols ; j ++ ) { sum += a [ j ] * x [ j ]; } y [ i ] = sum ; }         
                         

                     
             
                                 
             
            
    
      

カーネル関数matvec は、各呼び出しにおいて、行列Aの 1 行とベクトルxのドット積を計算します。

$${\displaystyle y_{i}=a_{i,:}\cdot x=\sum _{j}a_{i,j}x_{j}.}$$

これを完全な行列とベクトルの乗算に拡張するために、OpenCL ランタイムはカーネルを行列の行にマッピングします。ホスト側では、 clEnqueueNDRangeKernel関数がこれを実行します。この関数は、実行するカーネル、その引数、および行列Aの行数に対応する作業項目の数を引数として受け取ります。

この例では、高速フーリエ変換（FFT）実装をロードして実行します。実装を以下に示します。^[22]コードは、OpenCLライブラリに最初に利用可能なグラフィックカードを要求し、（グラフィックカードの観点から）読み取りと書き込み用のメモリバッファを作成し、 FFTカーネルをJITコンパイルし、最後にカーネルを非同期で実行します。この例では、変換の結果は読み取られません。

#include <stdio.h> 
#include <time.h> 
#include "CL/opencl.h" 

#定義 NUM_ENTRIES 1024

int main () // (int argc, const char* argv[]) の続きを読む  
{
	// 定数
	// カーネルのソースコードは文字列として表現されます
	// ファイル「fft1D_1024_kernel_src.cl」内にあります。詳細については、次のリストを参照してください。
	定数char *カーネルソース=   
		#include "fft1D_1024_kernel_src.cl" 
			;

	// 利用可能なGPUを検索
	定数cl_uint数値= 1 ;    
	clGetDeviceIDs ( NULL 、CL_DEVICE_TYPE_GPU 、0 、NULL 、( cl_uint * ) & num );    

	cl_device_idデバイス[ 1 ]; 
	clGetDeviceIDs ( NULL 、CL_DEVICE_TYPE_GPU 、num 、デバイス、NULL );    

	// GPUデバイスで計算コンテキストを作成する
	cl_contextコンテキスト= clCreateContextFromType ( NULL 、CL_DEVICE_TYPE_GPU 、NULL 、NULL 、NULL );       

	// コマンドキューを作成する
	clGetDeviceIDs ( NULL 、CL_DEVICE_TYPE_DEFAULT 、1 、デバイス、NULL );    
	cl_command_queueキュー= clCreateCommandQueue (コンテキスト、デバイス[ 0 ]、0 、NULL );      

	// バッファメモリオブジェクトを割り当てる
	cl_mem memobjs [] = { clCreateBuffer (コンテキスト、CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR 、sizeof ( float ) * 2 * NUM_ENTRIES 、NULL 、NULL )、              
						 clCreateBuffer (コンテキスト、CL_MEM_READ_WRITE 、sizeof ( float ) * 2 * NUM_ENTRIES 、NULL 、NULL ) };         

	// 計算プログラムを作成する
	// 定数char* fft1D_1024_kernel_src[1] = { };
	cl_programプログラム= clCreateProgramWithSource (コンテキスト, 1 , ( const char ** ) & KernelSource , NULL , NULL );          

	// 計算プログラム実行ファイルをビルドする
	clBuildProgram (プログラム、0 、NULL 、NULL 、NULL 、NULL );     

	// 計算カーネルを作成する
	cl_kernelカーネル= clCreateKernel (プログラム、"fft1D_1024" 、NULL );     

	// 引数の値を設定する

	size_tローカルワークサイズ[ 1 ] = { 256 };     

	clSetKernelArg (カーネル、0 、sizeof ( cl_mem )、( void * ) & memobjs [ 0 ] );    
	clSetKernelArg (カーネル、1 、sizeof ( cl_mem )、( void * ) & memobjs [ 1 ] );    
	clSetKernelArg (カーネル、2 、sizeof ( float ) * ( local_work_size [ 0 ] + 1 ) * 16 、NULL );       
	clSetKernelArg (カーネル、3 、sizeof ( float ) * ( local_work_size [ 0 ] + 1 ) * 16 、NULL );       

	// 作業項目の寸法を持つND範囲オブジェクトを作成し、カーネルを実行する
	size_tグローバルワークサイズ[ 1 ] = { 256 };     
	
	グローバルワークサイズ[ 0 ] =エントリ数;  
	local_work_size [ 0 ] = 64 ; //Nvidia: 192 または 256   
	clEnqueueNDRangeKernel (キュー、カーネル、1 、NULL 、グローバル作業サイズ、ローカル作業サイズ、0 、NULL 、NULL );        
}

ファイル「fft1D_1024_kernel_src.cl」内の実際の計算（「G80アーキテクチャへのFFTのフィッティング」に基づく）: ^[23]

R "( 
// このカーネルは長さ 1024 の FFT を計算します。長さ 1024 の FFT は、基数 16 の関数、別の基数 16 の関数、そして基数 4 の関数の呼び出しに分解されます。  
  

  __kernel void fft1D_1024 ( __global float2 * in 、__global float2 * out 、__local float * sMemx 、__local float * sMemy ) { int tid = get_local_id ( 0 ); int blockIdx = get_group_id ( 0 ) * 1024 + tid ; float2データ[ 16 ];        
                                
       
           
     

    // グローバル メモリとの間のデータの開始インデックス
in = in + blockIdx ; out = out + blockIdx ;              

    globalLoads ( data , in , 64 ); // 統合されたグローバル読み取りfftRadix16Pass ( data ); // インプレース radix-16 パスtwiddleFactorMul ( data , tid , 1024 , 0 );   
          
       

    // ローカル メモリを使用したローカル シャッフル
localShuffle ( data , sMemx , sMemy , tid , ((( tid & 15 ) * 65 ) + ( tid >> 4 ))); fftRadix16Pass (データ); // インプレース radix-16 はtwiddleFactorMul ( data , tid , 64 , 4 ); // 回転因子の乗算                
                   
        

    localShuffle ( data , sMemx , sMemy , tid , ((( tid >> 4 ) * 64 ) + ( tid & 15 )));            

    // 4 つの基数 4 関数呼び出し
fftRadix4Pass ( data ); // 基数 4 関数番号 1 fftRadix4Pass ( data + 4 ); // 基数 4 関数番号 2 fftRadix4Pass ( data + 8 ); // 基数 4 関数番号 3 fftRadix4Pass ( data + 12 ); // 基数 4 関数番号 4          
        
        
       

    // 結合されたグローバル書き込み
globalStores ( data , out , 64 ); } ) "      
  

OpenCL FFTの完全なオープンソース実装はAppleのウェブサイトで見つけることができます。^[24]

2020年にクロノスは、従来のOpenCL C機能と組み合わせてC++17の機能を提供する、コミュニティ主導のOpenCLプログラミング言語^{[26]のC++への移行を発表しました[25]}。この言語では、OpenCL Cとの下位互換性を維持しながら、標準C++の豊富な言語機能を活用できます。これにより、OpenCLカーネルコード開発者は、使い慣れたプログラミングフローやツールを引き続き使用できるだけでなく、OpenCL Cで利用可能な既存の拡張機能やライブラリを活用できるため、C++機能へのスムーズな移行パスが開かれます。

^{言語セマンティクスは、クロノスグループがホストするOpenCL-Docs [27]}リポジトリのリリースで公開されているドキュメントに記載されていますが、現在クロノスグループによって承認されていません。OpenCL用C++言語はスタンドアロンドキュメントには記載されておらず、C++とOpenCL Cの仕様に基づいています。オープンソースのClangコンパイラは、リリース9以降、OpenCL用C++をサポートしています。^[28]

C++ for OpenCLはもともとClangコンパイラー拡張として開発され、リリース9で登場しました。^[29] OpenCL Cと密結合されており、Clang固有の機能は含まれていなかったため、そのドキュメントは他の仕様やリファレンスカードのソースとともに、Khronos GroupのOpenCL-Docsリポジトリ^{[27]に再ホストされています。C++ for OpenCLバージョン1.0を説明するこのドキュメントの最初の公式リリースは、2020年12月に公開されました。 [30]} C++ for OpenCL 1.0にはC++17の機能が含まれており、OpenCL C 2.0と下位互換性があります。2021年12月には、OpenCL 3.0標準と完全に互換性のある新しい暫定的なC++ for OpenCLバージョン2021がリリースされました。^[31]最新のC++ for OpenCLドキュメントの作業中のドラフトは、KhronosのWebサイトでご覧いただけます。^[32]

C++ for OpenCL は、ネストされた並列処理とブロックを除いて、OpenCL C のほとんどの機能 (構文的および意味的) をサポートしています。^[33]ただし、サポートされている機能の中には、C++ と C の意味の違いに主に関連する小さな違いがあります。たとえば、C++ は暗黙の型変換に対してより厳格であり、restrict型修飾子をサポートしていません。^[33]次の C++ 機能は、C++ for OpenCL ではサポートされていません: 仮想関数、dynamic_cast演算子、非配置new / delete演算子、例外、メンバー関数へのポインター、関数への参照、C++ 標準ライブラリ。^{[33] C++ for OpenCL は、OpenCL C の個別のメモリ領域 (}アドレス空間)の概念を、関数キャスト、テンプレート、クラス メンバー、参照、ラムダ関数、および演算子などの C++ 機能に拡張します。C++ 機能のほとんど (オーバーロードやテンプレート、パラメーター型での任意のクラス レイアウトなど) は、カーネル関数では使用できません。^[33]

次のコード スニペットは、 C++ 機能の便利な使用により、複素数演算を備えたカーネルを OpenCL 言語の C++ で実装する方法を示しています。

// クラス Complex を定義します。このクラスは、 T に double、float、half などの異なる型が使用されている場合に、さまざまな精度
で複素数計算を実行できます。 template < typename T > class complex_t { T m_re ; // 実数部。T m_im ; // 虚数部。 
  
      
      

public : 
complex_t ( T re , T im ) : m_re { re }, m_im { im } {}; // 複素数乗算の演算子を定義します。complex_t operator * ( const complex_t & other ) const { return { m_re * other . m_re - m_im * other . m_im , m_re * other . m_im + m_im * other . m_re }; } T get_re () const { return m_re ; } T get_im () const { return m_im ; } };          
    
        
    
                
                      
    
          
          



//入力バッファ
から読み取った複素数の乗算を計算し、計算結果を出力バッファに格納するヘルパー関数。 template < typename T > void compute_helper ( __global T * in , __global T * out ) { auto idx = get_global_id ( 0 ); // すべての作業項目は、入力バッファから 4 つの連続する項目を使用します// - 各複素数につき 2 つ。auto offset = idx * 4 ; auto num1 = complex_t { in [ offset ], in [ offset + 1 ]}; auto num2 = complex_t { in [ offset + 2 ], in [ offset + 3 ]}; // 複素数の乗算を実行します。auto res = num1 * num2 ; // すべての作業項目は、出力バッファに 2 つの連続する項目を書き込みます。out [ idx * 2 ] = res . get_re (); out [ idx * 2 + 1 ] = res . get_im (); } 
       
           
    
    
         
          
            
    
         
    
        
          


// このカーネルは単精度の複素数乗算に使用されます。
__kernel void compute_sp ( __global float * in , __global float * out ) { compute_helper ( in , out ); }        
     


#ifdef cl_khr_fp16 // このカーネルは
、デバイスでサポートされている場合に、半精度の複素数乗算に使用されます。 
#pragma OPENCL EXTENSION cl_khr_fp16: enable __kernel void compute_hp ( __global half * in , __global half * out ) { compute_helper ( in , out ); } #endif

        
      

OpenCL 言語の C++ は、OpenCL C 言語と同じアプリケーションやライブラリに、同じ方法で使用できます。C++ 言語の豊富な機能により、OpenCL 言語の C++ で記述されたアプリケーションは、OpenCL C で記述されたアプリケーションよりも複雑な機能をより便利に表現できます。特に、 C++ の汎用プログラミングパラダイムは、ライブラリ開発者にとって非常に魅力的です。

OpenCL用C++ソースは、 cl_ext_cxx_for_opencl拡張機能をサポートするOpenCLドライバーでコンパイルできます。^[34] Armは2020年12月にこの拡張機能のサポートを発表しました。^{[35]ただし、OpenCLデバイスで高速化されるアルゴリズムの複雑さが増しているため、Clang [36]}などのスタンドアロンコンパイラを使用して、OpenCLカーネル用のC++をオフラインで実行可能バイナリ形式またはSPIR-Vなどのポータブルバイナリ形式にコンパイルするアプリケーションが増えると予想されます。^[37]このような実行可能ファイルは、専用のOpenCL APIを使用してOpenCLアプリケーションの実行中にロードできます。^[38]

OpenCL 1.0 用の C++ ソースからコンパイルされたバイナリは、OpenCL 2.0 準拠デバイスで実行できます。このようなカーネル ソースで使用される言語機能に応じて、以前の OpenCL バージョンまたは OpenCL 3.0 をサポートするデバイスでも実行できます。

OpenCLドライバーとは別に、OpenCL用にC++で書かれたカーネルは、OpenCL Cカーネルと同じように clspv ^[39]コンパイラとclvk ^{[40]ランタイムレイヤーを使用してVulkanデバイス上で実行するためにコンパイルできます。}

OpenCL用のC++は、ドキュメントに記載されている貢献者のコミュニティによって開発されたオープン言語です。^[32] 言語のセマンティック定義やオープンソースツールのサポートへの新しい貢献は、主な設計理念に沿っており、経験豊富な貢献者によってレビューされ承認され次第、興味のある人から受け入れられます。^[19]

OpenCLは当初、商標権を持つApple Inc.によって開発され、 AMD、IBM、Qualcomm、Intel、Nvidiaの技術チームと共同で最初の提案に洗練されました。Appleはこの最初の提案をKhronos Groupに提出しました。2008年6月16日、CPU、GPU、組み込みプロセッサ、ソフトウェア企業の代表者でKhronos Compute Working Groupが結成されました^[41]。このグループは5か月間作業し、2008年11月18日までにOpenCL 1.0の仕様の技術的詳細を完成させました。^[42]この技術仕様はKhronosメンバーによってレビューされ、2008年12月8日に公開が承認されました。^[43]

OpenCL 1.0は2009年8月28日にMac OS X Snow Leopardでリリースされました。Appleのプレスリリースによると: ^[44]

Snow Leopard では、Open Computing Language (OpenCL) によって最新のハードウェアのサポートがさらに拡張され、これまではグラフィックス アプリケーションでしか利用できなかった膨大なギガフロップスの GPU コンピューティング パワーをあらゆるアプリケーションで利用できるようになります。OpenCL は C プログラミング言語に基づいており、オープン スタンダードとして提案されています。

AMDは、 Streamフレームワークで、現在は非推奨となっているClose to Metalの代わりにOpenCLをサポートすることを決定しました。^{[45] [46]} RapidMindは、1つのインターフェイスで複数のベンダーのGPUをサポートするために、開発プラットフォームでOpenCLを採用すると発表しました。^[47] 2008年12月9日、Nvidiaは、GPUコンピューティングツールキットにOpenCL 1.0仕様の完全なサポートを追加する意向を発表しました。^[48] 2009年10月30日、IBMはXLコンパイラの一部として最初のOpenCL実装をリリースしました。^[49]

OpenCLでは、通常のCPUに比べてグラフィックカードで1000倍の計算速度を実現できます。 [ ⁵⁰ ] 倍精度や半精度の演算など、OpenCLの次期バージョンの重要な機能のいくつかは、1.0ではオプションです。^[51]

OpenCL 1.1は2010年6月14日にクロノスグループによって承認され、^[52]並列プログラミングの柔軟性、機能性、パフォーマンスを向上させる重要な機能が追加されました。

• 3 成分ベクトルや追加の画像形式を含む新しいデータ タイプ。
• 複数のホスト スレッドからのコマンドを処理し、複数のデバイス間でバッファーを処理します。
• 1D、2D、または 3D の長方形領域の読み取り、書き込み、コピーを含む、バッファの領域に対する操作。
• コマンド実行を駆動および制御するためのイベントの使用が強化されました。
• 整数クランプ、シャッフル、非同期ストライド コピーなどの追加の OpenCL 組み込み C 関数。
• OpenCL と OpenGL のイベントをリンクすることで、イメージとバッファを効率的に共有し、OpenGL の相互運用性が向上しました。

2011年11月15日、クロノスグループはOpenCL 1.2仕様^[53]を発表しました。この仕様では、並列プログラミングのパフォーマンスと機能の面で、以前のバージョンに比べて大幅な機能が追加されました。最も注目すべき機能は次のとおりです。

• デバイスのパーティショニング: デバイスをサブデバイスに分割して、作業割り当てを個々のコンピューティング ユニットに割り当てる機能。これは、デバイスの領域を予約して、時間的に重要なタスクの待ち時間を減らすのに役立ちます。
• オブジェクトの個別のコンパイルとリンク: OpenCL を外部ライブラリにコンパイルして他のプログラムに組み込む機能。
• 強化された画像サポート (オプション): 1.2 では、1D 画像と 1D/2D 画像配列のサポートが追加されました。さらに、OpenGL 共有拡張機能により、OpenGL 1D テクスチャと 1D/2D テクスチャ配列を使用して OpenCL 画像を作成できるようになりました。
• 組み込みカーネル: 特定の独自機能を備えたカスタム デバイスが、OpenCL フレームワークにさらに密接に統合されました。カーネルを呼び出して、基盤となるハードウェアの特殊な側面やプログラム不可能な側面を使用できます。例としては、ビデオ エンコード/デコードやデジタル信号プロセッサなどがあります。
• DirectX 機能: DX9 メディア サーフェス共有により、OpenCL と DX9 またはDXVAメディア サーフェス間の効率的な共有が可能になります。同様に、DX11 では、OpenCL と DX11 サーフェス間のシームレスな共有が有効になります。
• 単精度浮動小数点演算のIEEE 754準拠を強制する機能: OpenCL はデフォルトで、単精度バージョンの除算、逆数、平方根演算が IEEE 754 で要求される正しく丸められた値よりも精度が低くなることを許可します。 ^[54]プログラマーが「-cl-fp32-correctly-rounded-divide-sqrt」コマンドライン引数をコンパイラーに渡すと、OpenCL 実装がこれをサポートしている場合は、これら 3 つの演算が IEEE 754 要件に従って計算され、OpenCL 実装がこれらの演算を IEEE 754 仕様で定義されている正しく丸められた値に計算することをサポートしていない場合は、コンパイルに失敗します。^[54]この機能は、OpenCL 実装にクエリを実行して、これらの演算を IEEE 754 の精度で実行できるかどうかを判断する機能によって補完されます。^[54]

2013年11月18日、クロノスグループは最終的なOpenCL 2.0仕様の承認と一般公開を発表しました。^[55] OpenCL 2.0の更新と追加には以下が含まれます。

• 共有仮想メモリ
• ネストされた並列処理
• 汎用アドレス空間
• 画像（オプション、3D画像を含む）
• C11原子
• パイプ
• Androidにインストール可能なクライアント ドライバー拡張機能
• オプションの cl_khr_fp16 拡張により半精度が拡張されました
• cl_double: 倍精度IEEE 754（オプション）

OpenCL 2.1暫定仕様の承認とリリースは、2015年3月3日にサンフランシスコで開催されたゲーム開発者会議で発表されました。2015年11月16日にリリースされました。^[56] C++14のサブセットに基づくOpenCL C++カーネル言語が導入されましたが、既存のOpenCL Cカーネル言語のサポートは維持されています。VulkanとOpenCL 2.1は中間表現としてSPIR-Vを共有しており、高水準言語フロントエンドが共通のコンパイルターゲットを共有できるようにしています。OpenCL APIの更新には以下が含まれます。

• 追加のサブグループ機能
• カーネルオブジェクトと状態のコピー
• 低遅延デバイスタイマークエリ
• ランタイムによる SPIR-V コードの取り込み
• キューの実行優先度のヒント
• ホストからのゼロサイズのディスパッチ

AMD、ARM、Intel、HPC、YetiWareはOpenCL 2.1のサポートを宣言している。^{[57] [58]}

OpenCL 2.2では、OpenCL C++カーネル言語がコア仕様に導入され、並列プログラミングの生産性が大幅に向上しました。^{[59] [60] [61]} 2017年5月16日にリリースされました。^[62] 2018年5月にバグ修正を含むメンテナンスアップデートがリリースされました。^[63]

• OpenCL C++ カーネル言語はC++14標準の静的サブセットであり、クラス、テンプレート、ラムダ式、関数オーバーロード、およびジェネリックプログラミングとメタプログラミングのための他の多くの構成要素が含まれています。
• OpenCL C++ カーネル言語を完全にサポートする新しい Khronos SPIR-V 1.1 中間言語を使用します。
• OpenCL ライブラリ関数は、C++ 言語を使用して、アトミック、イテレータ、イメージ、サンプラー、パイプ、デバイス キューの組み込み型やアドレス空間などの機能にアクセスする際に、安全性を高め、未定義の動作を削減できるようになりました。
• パイプ ストレージは OpenCL 2.2 の新しいデバイス側タイプであり、コンパイル時に接続のサイズとタイプを把握することでカーネル間の効率的なデバイス スコープ通信を可能にし、FPGA 実装に役立ちます。
• OpenCL 2.2 には、生成されたコードの最適化を強化する機能も含まれています。アプリケーションは SPIR-V コンパイル時に特殊化定数の値を提供でき、新しいクエリはプログラム スコープのグローバル オブジェクトの重要なコンストラクターとデストラクターを検出でき、プログラムのリリース時にユーザー コールバックを設定できます。
• OpenCL 2.0 対応のハードウェアであればどこでも実行できます (ドライバーの更新のみが必要です)。

OpenCL 3.0仕様は、2020年4月からプレビュー段階を経て、2020年9月30日にリリースされました。OpenCL 1.2の機能は必須のベースラインとなり、OpenCL 2.xとOpenCL 3.0の機能はすべてオプションとなりました。この仕様ではOpenCL C言語が保持され、OpenCL C++カーネル言語は非推奨となり、C++17のサブセットとSPIR-V中間コードを実装するClang / LLVMコンパイラに基づくC++ for OpenCL言語^{[19]に置き換えられました。 [64] [65] [66]} 一部のKhronos openCL拡張機能を備えたC++ for OpenCLバージョン3.0.7がIWOCL 21で発表されました。^[67]実際のバージョンは、いくつかの新しい拡張機能と修正を加えた3.0.11です。NVIDIAは、Khronos OpenCLワーキンググループと緊密に協力し、セマフォとメモリ共有によりVulkan Interopを改善しました。^[68]最後のマイナーアップデートは3.0.14で、バグ修正と複数のデバイス向けの新しい拡張機能が追加されました。^[69]

OpenCL 2.2 をリリースする際、クロノス グループは、OpenCL が可能な限りVulkanと統合し、両方の API で OpenCL ソフトウェアの展開の柔軟性を実現すると発表しました。^{[70] [71]}これは、Adobe の Premiere Rush が clspv ^[39]オープンソース コンパイラを使用して大量の OpenCL C カーネル コードをコンパイルし、Android に展開するために Vulkan ランタイムで実行することで実証されました。^[72] OpenCL には Vulkan とは独立した将来を見据えたロードマップがあり、「OpenCL Next」が開発中であり、2020 年のリリースを目標としています。OpenCL Next は、Vulkan / OpenCL Interop、スクラッチパッド メモリ管理、拡張サブグループ、SPIR-V 1.4 取り込み、SPIR-V 拡張デバッグ情報などの拡張機能を統合する可能性があります。OpenCL は、複数のアクセラレータ タイプでの展開の柔軟性のために、Vulkan のようなローダーとレイヤー、および「柔軟なプロファイル」も検討しています。^[73]

OpenCL は、一連のヘッダーと実行時にロードされる共有オブジェクトで構成されています。ランタイムがサポートする必要があるすべてのベンダー クラスについて、インストール可能なクライアント ドライバー (ICD) をプラットフォームにインストールする必要があります。つまり、たとえば Linux プラットフォームで Nvidia デバイスをサポートするには、OpenCL ランタイム (ICD ローダー) がベンダーの ICD を見つけて呼び出しを適切にリダイレクトできるように、Nvidia ICD をインストールする必要があります。標準の OpenCL ヘッダーはコンシューマー アプリケーションによって使用され、各関数の呼び出しは、ICD を使用して OpenCL ランタイムによって適切なドライバーにプロキシされます。各ベンダーは、各 OpenCL 呼び出しをドライバーに実装する必要があります。^[74]

Apple、^[75]、 Nvidia、^[76]、 ROCm、RapidMind ^[77]、Gallium3D ^{[78]によるOpenCLの実装はすべて}LLVMコンパイラ技術に基づいており、フロントエンドとして Clangコンパイラを使用しています。

MESA ガリウム コンピューティング

多数のプラットフォーム向けのOpenCLの実装（実際の1.1は不完全で、大部分はAMD Radeon GCNで完了）は、Gallium Compute Project ^{[79]の一部として維持されています。これは、複数のプラットフォームをサポートするために}Mesaプロジェクトの成果に基づいています。以前はCLOVERとして知られていました。^[80]実際の開発：実際のLLVMとCLANGで不完全なフレームワークを実行するためのサポートがほとんどで、17.3ではfp16などのいくつかの新機能が追加されました。 ^[81] AMDとNvidia向けの完全なOpenCL 1.0、1.1、1.2をターゲットとしています。新しい基本開発は、 Clover向けにもRed HatによってSPIR-Vで行われています。^{[82] [83]}新しいターゲットは、OpenCL 1.2を完全にサポートするモジュラーOpenCL 3.0です。実際の状態はMesamatrixで利用できます。ここではイメージのサポートが開発の焦点となっています。

RustiCLは、Cの代わりにRustを使用したGalliumコンピューティングの新しい実装です。Mesa 22.2では、openCL 3.0サポートとDarktableなどのプログラム用の画像拡張実装を備えた実験的な実装が利用できます。 ^[84] Intel Xe（Arc）とAMD GCN+は、Mesa 22.3以降でサポートされています。AMD R600とNvidia Kepler+もハードウェアサポートの対象です。^{[85] [86] [87]} RustiCLは、LuxmarkベンチマークでRadeon RX 6700 XTハードウェアを搭載したAMD ROCMよりも優れたパフォーマンスを発揮します。^[88] Mesa 23.1は公式のRustiCLをサポートしています。Mesa 23.2では、重要なfp64のサポートは実験レベルです。

MicrosoftのWindows 11 on Armでは、 Mesa Gallium経由でDirectX 12上にオープンソースのOpenCL実装であるCLon12を介してOpenCL 1.2のサポートが追加されました。^{[89] [90] [91]}

ベニエ

2013年にIntelによるIvy Bridge +ハードウェアの実装がリリースされました。^[92] Intelの中国チームによるこのソフトウェアは、AMDとRed Hatの開発者、^[93] PhoronixのMichael Larabelから批判を集めています。^[94]実際のバージョン1.3.2は、OpenCL 1.2を完全にサポートしています（Ivy Bridge以上）、およびSkylake以降ではOpenCL 2.0がオプションです。^{[95] [96]} BeignetにAndroidのサポートが追加されました。、^[97]実際の開発目標：1.2と2.0のみのサポート、OpenCL 2.1、2.2、3.0への道はNEOにあります。

ネオ

2018年にリリースされたGen. 8 Broadwell + Gen. 9ハードウェア向けのIntelによる実装。^[98]このドライバーは、サポートされているプラ​​ットフォーム（古い6.genからHaswellまでではない）のBeignet実装を置き換えます。NEOは、CoreプラットフォームでOpenCL 2.1のサポートを提供し、AtomプラットフォームでOpenCL 1.2を提供します。^[99] 2020年現在は、Graphic Gen 11 Ice LakeとGen 12 Tiger Lakeもサポートされています。新しいOpenCL 3.0は、バージョン20.41以降でAlder Lake、Tiger LakeからBroadwellまで利用できます。これには、オプションのOpenCL 2.0、2.1の完全な機能、および2.2の一部の機能が含まれています。

ROCm

AMDのGPUOpenの一環として作成されたROCm （Radeon Open Compute）は、OpenCL 1.2上に構築され、2.0の言語サポートを備えたオープンソースのLinuxプロジェクトです。このシステムは、すべての最新のAMD CPUおよびAPU（実際には部分的にGFX 7、GFX 8および9）、およびIntel Gen7.5+ CPU（PCI 3.0のみ）と互換性があります。^{[100] [101]}バージョン1.9では、サポートがいくつかの点でPCIe 2.0およびアトミックなしのハードウェアに実験的に拡張されています。実際の作業の概要はXDC2018で行われます。^{[102] [103]} ROCmバージョン2.0は完全なOpenCL 2.0をサポートしていますが、いくつかのエラーと制限がtodoリストにあります。^{[104] [105]}バージョン3.3は細部が改善されています。^[106]バージョン3.5はOpenCL 2.2をサポートしています。^[107]バージョン3.10では、改良と新しいAPIが追加されました。^[108] SC20で発表されたROCm 4.0では、AMD Compute Card Instinct MI 100がサポートされています。 ^[109] 5.5.1以前の実際のドキュメントはGitHubで入手できます。^{[110] [111] [112]} OpenCL 3.0が利用可能です。RocM 5.5.x+はGFX 9 Vega以降のみをサポートしているため、代替案としては古いRocMリリースか、将来的には古いハードウェア用のRustiCLが考えられます。

ポクル

CPU と一部の GPU ( CUDAおよびHSA経由)をサポートするポータブル実装。ClangとLLVM上に構築されています。^[113]バージョン 1.0 では、OpenCL 1.2 がいくつかの 2.x 機能とともにほぼ完全に実装されました。^{[114]バージョン 1.2 は}、マイルストーン 1.2 ですべてのクローズされたチケットを備えた LLVM/CLANG 6.0、7.0、および完全な OpenCL 1.2 サポートを備えています。^{[114] [115]} OpenCL 2.0 はほぼ完全に実装されています。^[116]バージョン 1.3 は Mac OS X をサポートしています。^{[117]バージョン 1.4 には、} LLVM 8.0 および 9.0のサポートが含まれています。 ^[118]バージョン 1.5 は LLVM/Clang 10 のサポートを実装しています。^[119]バージョン 1.6 は LLVM/Clang 11 のサポートと CUDA アクセラレーションを実装しています。^[120]実際のターゲットは、OpenCL 2.x、OpenCL 3.0の完全サポートとパフォーマンスの向上です。POCL 1.6は、Intelコンピューティングランタイムと同じレベルの手動最適化を備えています。^[121]バージョン1.7は、LLVM/Clang 12のサポートといくつかの新しいOpenCL 3.0機能を実装しています。^[122]バージョン1.8は、LLVM/Clang 13のサポートを実装しています。^[123]バージョン3.0は、最小レベルのOpenCL 3.0とLLVM/Clang 14を実装しています。 ^[124]バージョン3.1は、LLVM/Clang 15と改善されたSpir-Vサポートで動作します。^[125]

シャムロック

Mesa CloverのARM版でOpenCL 1.2を完全にサポートしているが、^{[126] [127]} 2.0向けの開発は実際には行われていない。

無料OCL

より信頼性の高いプラットフォームを作成するために外部コンパイラを実装したOpenCL 1.2のCPUに重点を置いた実装。 ^[128]実際の開発は行われていません。

商工労働省

NUDT研究者によるMatrix-2000用のPOCLに基づくOpenCL実装は2018年にリリースされました。Matrix-2000アーキテクチャは、TianHe-2スーパーコンピューターのIntel Xeon Phiアクセラレーターを置き換えるように設計されています。このプログラミングフレームワークはLLVM v5.0上に構築されており、POCLのコードの一部も再利用しています。ハードウェアの潜在能力を解き放つために、デバイスランタイムはプッシュベースのタスクディスパッチ戦略を使用し、カーネルアトミックのパフォーマンスが大幅に向上しています。このフレームワークはTH-2Aシステムに導入されており、一般に公開されています。^[129]ソフトウェアの一部は次にPOCLを改善するために移植される予定です。^[114]

VC4CL

Raspberry Piモデル4以前に使用されていたVideoCore IV（BCM2763）プロセッサ用のOpenCL 1.2実装。^[130]

• 2008 年 6 月: Apple の WWDC カンファレンス中に、 Mac OS X Snow Leopardの初期ベータ版が参加者に提供されました。これには OpenCL の最初のベータ実装が含まれていました。これは、最終バージョン 1.0 仕様が 2008 年後半に承認される約 6 か月前のことでした。また、2 つのデモも披露されました。1 つは、レンダリングされた 8×8 画面のグリッドで、それぞれにエミュレートされた Apple II マシンの画面が表示されていました。合計 64 の独立したインスタンスがあり、それぞれが有名な空手ゲームを実行していました。これは、CPU でのタスクの並列処理を示していました。もう 1 つのデモは、 Mac Pro の GPU で実行されるN体シミュレーションで、データ並列タスクでした。
• 2008年12月10日: AMDとNvidiaは、 SIGGRAPH Asia 2008で75分間のプレゼンテーションを行い、初の公開OpenCLデモを開催した。AMDはCPUアクセラレーションによるOpenCLデモを行い、1つ以上のコアでのOpenCLのスケーラビリティを説明した。一方、NvidiaはGPUアクセラレーションによるデモを行った。^{[131] [132]}
• 2009年3月16日：第4回マルチコアエキスポで、イマジネーションテクノロジーズは同社初のOpenCLサポートを搭載したGPUであるPowerVR SGX543MPを発表しました。 ^[133]
• 2009年3月26日: GDC 2009で、AMDとHavokはATI Radeon HD 4000シリーズGPU上でOpenCLアクセラレーションHavok Clothの最初の実用的な実装を実演しました。 ^[134]
• 2009年4月20日: NvidiaはOpenCL早期アクセスプログラムに参加している開発者向けにOpenCLドライバーとSDKのリリースを発表しました。 ^[135]
• 2009年8月5日: AMDはATI Stream SDK v2.0ベータプログラムの一環として、OpenCLプラットフォーム用の最初の開発ツールを発表しました。 ^[136]
• 2009年8月28日: AppleはOpenCLの完全な実装を含むMac OS X Snow Leopardをリリースした。 ^[137]
• 2009 年 9 月 28 日: Nvidia は独自の OpenCL ドライバーと SDK 実装をリリースしました。
• 2009年10月13日: AMDはATI Stream SDK 2.0の4番目のベータ版をリリースしました。これはR700 / HD 5000 GPUとSSE3対応CPUの両方で完全なOpenCL実装を提供します。SDKはLinuxとWindowsの両方で利用可能です。^[138]
• 2009 年 11 月 26 日: Nvidia は OpenCL 1.0 (rev 48) 用のドライバーをリリースしました。
• 2009年10月27日：S3はネイティブOpenCL 1.0をサポートする最初の製品であるChrome 5400E組み込みグラフィックプロセッサをリリースしました。^[139]
• 2009年12月10日: VIAはOpenCL 1.0をサポートする最初の製品であるVN1000チップセットに搭載されたChromotionHD 2.0ビデオプロセッサをリリースしました。^[140]
• 2009年12月21日: AMDはATI Stream SDK 2.0の製品版をリリースしました。^{[141]これは}HD 5000 GPUのOpenCL 1.0サポートとR700 GPUのベータサポートを提供します。
• 2010年6月1日: ZiiLABSは、ハンドヘルド、組み込み、デジタルホーム製品向けのZMSプロセッサの最初のOpenCL実装の詳細を発表しました。^[142]
• 2010年6月30日: IBMはOpenCL 1.0の完全準拠バージョンをリリースしました。^[4]
• 2010年9月13日: IntelはSandy Bridgeチップアーキテクチャ向けの初のOpenCL実装の詳細を発表しました。Sandy BridgeはIntelの最新のグラフィックスチップ技術を中央処理装置に直接統合します。^[143]
• 2010年11月15日：Wolfram ResearchはOpenCLLink ^{[144]パッケージを搭載した}Mathematica 8をリリースした。
• 2011年3月3日: Khronos Groupは、OpenCLへのJavaScriptバインディングの定義を検討するためのWebCLワーキンググループの結成を発表しました。これにより、 WebブラウザからGPUとマルチコアCPUの並列処理を活用できるようになります。^{[145] [146]}
• 2011年3月31日: IBMはOpenCL 1.1の完全準拠バージョンをリリースした。^{[4] [147]}
• 2011年4月25日: IBMはx86アーキテクチャ上のLinux向けOpenCL Common Runtime v0.1をリリースした。^[148]
• 2011年5月4日: Nokia ResearchはFirefoxウェブブラウザ用のオープンソースのWebCL拡張機能をリリースし、OpenCLへのJavaScriptバインディングを提供しました。^[149]
• 2011年7月1日：サムスン電子はWebKit用のWebCLのオープンソースプロトタイプ実装をリリースし、OpenCLへのJavaScriptバインディングを提供した。^[150]
• 2011年8月8日: AMDは、技術とコンセプトの両面でATI Stream SDKに代わる、OpenCL駆動のAMD Accelerated Parallel Processing (APP)ソフトウェア開発キット(SDK)v2.5をリリースしました。 ^[151]
• 2011年12月12日：AMDはOpenCL 1.2のプレビューを含むAMD APP SDK v2.6 ^{[152]をリリースした。}
• 2012年2月27日:ポートランドグループは、マルチコアARM CPU用のPGI OpenCLコンパイラをリリースしました。^[153]
• 2012年4月17日: KhronosはWebCLのワーキングドラフトをリリースしました。^[154]
• 2013年5月6日：アルテラはAltera SDK for OpenCLバージョン13.0をリリースしました。^[155]これはOpenCL 1.0に準拠しています。^[156]
• 2013年11月18日：クロノスはOpenCL 2.0の仕様が完成したと発表した。^[157]
• 2014年3月19日: KhronosがWebCL 1.0仕様をリリース。^{[158] [159]}
• 2014年8月29日: IntelがOpenCL 2.0をサポートするHD Graphics 5300ドライバをリリース。^[160]
• 2014年9月25日: AMDはOpenCL 2.0ドライバを含むCatalyst 14.41 RC1をリリースしました。^[161]
• 2015年1月14日：ザイリンクス社はOpenCL、C、C++向けのSDAccel開発環境を発表し、クロノス準拠を達成した。^[162]
• 2015年4月13日: NvidiaはWHQLドライバーv350.12をリリースしました。これにはKepler以降のアーキテクチャに基づくGPUのOpenCL 1.2サポートが含まれています。^[163]ドライバー340+はTeslaとFermiのOpenCL 1.1をサポートしています。
• 2015年8月26日：AMDはOpenCL 2.0の完全なサポートとサンプルコーディングを含むAMD APP SDK v3.0 ^{[164]をリリースしました。}
• 2015年11月16日：クロノスはOpenCL 2.1の仕様が完成したと発表した。^[165]
• 2016年4月18日：クロノスはOpenCL 2.2の仕様が暫定的に完成したと発表した。^[60]
• 2016年11月3日: IntelがSDK 2016 r3でOpenCL 2.1のGen7+をサポート。^[166]
• 2017年2月17日: Nvidiaはドライバー378.66でOpenCL 2.0の評価サポートを開始しました。^{[167] [168] [169]}
• 2017年5月16日：クロノスは、OpenCL 2.2の仕様がSPIR-V 1.2で完成したと発表した。^[170]
• 2018年5月14日：クロノスはバグ修正と統合ヘッダーを含むOpenCL 2.2のメンテナンスアップデートを発表しました。^[63]
• 2020 年 4 月 27 日: Khronos が OpenCL 3.0 の暫定バージョンを発表しました。
• 2020 年 6 月 1 日: 新しい Tiger Lake 向けの OpenCL 3.0 を搭載した Intel NEO ランタイム。
• 2020年6月3日：AMDはOpenCL 2.2をサポートするRocM 3.5を発表しました。^[171]
• 2020 年 9 月 30 日: Khronos は、OpenCL 3.0 の仕様が確定したことを発表しました (CTS も利用可能)。
• 2020 年 10 月 16 日: Intel は NEO 20.41 で OpenCL 3.0 (オプションの OpenCL 2.x のほとんどを含む) をサポートすることを発表しました。
• 2021年4月6日：NvidiaはAmpereでOpenCL 3.0をサポートします。Maxwell以降のGPUもNvidiaドライバー465+でOpenCL 3.0をサポートします。^[172]
• 2022年8月20日：Intel Arc Alchemist GPU（Arc A380、A350M、A370M、A550M、A730M、A770M）はOpenCL 3.0に準拠しています。^[173]
• 2022 年 10 月 14 日: Arm Mali-G615 および Mali-G715-Immortalis は OpenCL 3.0 に準拠しました。^[173]
• 2022年11月11日：Rusticl OpenCLライブラリはOpenCL 3.0に準拠しました。^{[173] [174]}

2016 年現在、OpenCL はグラフィックス プロセッシング ユニット(GPU)、SIMD命令を備えたCPU 、 FPGA、Movidius Myriad 2、Adapteva Epiphany、DSP上で実行されます。

公式に準拠するには、実装がクロノス適合テストスイート（CTS）に合格する必要があり、その結果はクロノスアドプタープログラムに提出される。^[175]すべてのOpenCLバージョンのクロノスCTSコードは、2017年からオープンソースで利用可能となっている。^[176]

KhronosグループはOpenCL準拠製品の拡張リストを維持しています。^[4]

標準に準拠した実装はすべて、clinfoツールの1つを使用して照会できます（同じ名前で類似の機能セットを持つツールが複数あります）。^{[187] [188] [189]}

OpenCLサポートの製品とそのバージョンは次のとおりです。^[190]

OpenCL 1.2+を搭載したすべてのハードウェアが可能、OpenCL 2.xはオプションのみ、Khronos Test Suiteは2020-10から利用可能^{[191] [192]}

• (2020) Intel NEO Compute: 20.41+ (第12世代 Tiger Lake から Broadwell まで) (2.0 と 2.1 の完全サポートと 2.2 の一部を含む) ^[193]
• (2020) 最新の Intel Windows グラフィック ドライバーを搭載した Intel 第 6、第 7、第 8、第 9、第 10、第 11 世代プロセッサー ( Skylake、Kaby Lake、Coffee Lake、Comet Lake、Ice Lake、Tiger Lake )
• (2021) 最新の Intel Windows グラフィックス ドライバーを搭載した Intel 第 11 世代、第 12 世代プロセッサー ( Rocket Lake、Alder Lake )
• (2021) Arm Mali-G78、Mali-G310、Mali-G510、Mali-G610、Mali-G710、および Mali-G78AE。
• (2022)最新の Intel Windows グラフィックス ドライバーを搭載したIntel 第 13 世代プロセッサー ( Raptor Lake )
• (2022)最新の Intel Arc Windows グラフィックス ドライバーを搭載したIntel Arcディスクリート グラフィックス
• （2021）Nvidia Maxwell、Pascal、Volta、Turing、Ampere（Nvidiaグラフィックドライバー465+搭載）。^[172]
• (2022) Nvidia Ada Lovelaceと Nvidia グラフィック ドライバー 525+。
• (2022) Samsung Xclipse 920 GPU (AMD RDNA2 ベース)
• (2023) Intel 第 14 世代プロセッサ ( Raptor Lake ) 最新の Intel Windows グラフィックス ドライバーで更新
• (2023) Intel Core Ultra シリーズ 1 プロセッサ ( Meteor Lake ) と最新の Intel Windows グラフィックス ドライバー

まだありません: Khronos テスト スイートの準備が整い、すべてのハードウェアでドライバー アップデートにより 2.0 および 2.1 がサポートされる可能性があります

• Intel NEO Compute: 実際の製品に向けて作業中^[194]
• ROCm: 主にバージョン3.5+

• (2018+) Intel 第 5 世代および第 6 世代プロセッサ ( Broadwell、Skylake )へのサポートがバックポートされました。
• (2017+) Intel 第 7、第 8、第 9、第 10 世代プロセッサ ( Kaby Lake、Coffee Lake、Comet Lake、Ice Lake )
• Khronos: ドライバーアップデートにより、すべてのハードウェアで 2.0 のサポートが可能

• (2011+) AMD GCN GPU (HD 7700+/HD 8000/Rx 200/Rx 300/Rx 400/Rx 500/Rx 5000 シリーズ)、一部の GCN 1st Gen のみ 1.2 と一部の拡張機能
• (2013+) AMD GCN APU ( Jaguar、Steamroller、Puma、Excavator、Zenベース)
• (2014+) Intel 第 5 世代および第 6 世代プロセッサ ( Broadwell、Skylake )
• (2015+) Qualcomm Adreno 5xx シリーズ
• (2018+) Qualcomm Adreno 6xx シリーズ
• (2017+) Android 7.1 および Linux の ARM Mali (Bifrost) G51 および G71
• (2018+) ARM Mali (Bifrost) G31、G52、G72、G76
• (2017+) 不完全な評価サポート: Nvidia Kepler、Maxwell、Pascal、Volta、Turing GPU (GeForce 600、700、800、900、10 シリーズ、Quadro K、M、P シリーズ、Tesla K、M、P シリーズ)、ドライバー バージョン 378.66+

• (2011+) 一部の AMD GCN 1st Gen では、一部の OpenCL 2.0 機能は現在は利用できませんが、 Terascaleよりも多くの拡張機能があります。
• (2009+) AMD TeraScale 2 および 3 GPU (HD 5000、6000、7000 シリーズの RV8xx、RV9xx)
• (2011+) AMD TeraScale APU ( K10、Bobcat、Piledriverベース)
• (2012+) Nvidia Kepler、Maxwell、Pascal、Volta、Turing GPU (GeForce 600、700、800、900、10、16、20 シリーズ、Quadro K、M、P シリーズ、Tesla K、M、P シリーズ)
• (2012+) Intel 第 3 世代および第 4 世代プロセッサ ( Ivy Bridge、Haswell )
• (2013+) Qualcomm Adreno 4xx シリーズ
• (2013+) ARM マリミッドガルド第 3 世代 (T760)
• (2015+) ARM Mali Midgard 第4世代 (T8xx)

• (2008+) 一部の AMD TeraScale 1 GPU (HD4000 シリーズの RV7xx)
• (2008+) Nvidia Tesla、Fermi GPU (GeForce 8、9、100、200、300、400、500 シリーズ、Quadro シリーズ、または Tesla シリーズ ( TeslaまたはFermi GPU 搭載)
• (2011+) Qualcomm Adreno 3xx シリーズ
• (2012+) ARM Mali Midgard 第 1 世代および第 2 世代 (T-6xx、T720)

• 1.0専用ドライバーの後に、ほとんどが1.1と1.2に更新されました

OpenCL の重要な機能は、抽象化されたメモリと実行モデルによる移植性であり、プログラマーは他のプラットフォームへの直接移植性を放棄しない限り、Nvidia GPU のインライン並列スレッド実行(PTX) などのハードウェア固有のテクノロジを直接使用することはできません。準拠した実装であれば、どの OpenCL カーネルでも実行できます。

しかし、カーネルのパフォーマンスは必ずしもプラットフォーム間で移植可能というわけではありません。カーネルコードが適切に調整されている場合、既存の実装は競争力があることが示されており、自動チューニングはパフォーマンス移植性問題の解決策として提案されており、^[195]実験的な線形代数カーネルで「許容可能なレベルのパフォーマンス」が得られています。^[196]異なる動作を持つ複数のカーネルを含むアプリケーション全体の移植性も研究され、移植性には限られたトレードオフしか必要ないことが示されています。^[197]

2011年にデルフト大学で行われたCUDAプログラムとOpenCL Cへの直接的な変換を比較した研究では、CUDAはNvidia実装でOpenCLを最大30%上回るパフォーマンスを示した。研究者らは、OpenCLプログラムに手動の最適化を適用することで比較をより公平にすることができ、その場合「OpenCLがCUDAよりもパフォーマンスが劣る理由はない」と指摘した。パフォーマンスの違いは主にプログラミングモデル（特にメモリモデル）の違いと、CUDAとOpenCLのNVIDIAコンパイラの最適化の違いに起因すると考えられる。^[195]

D-Wave Systems Inc.の別の調査では、「OpenCLカーネルのパフォーマンスはCUDAのパフォーマンスよりも約13%から63%遅く、エンドツーエンドの時間は約16%から67%遅い」ことが判明しました。^[198]

OpenCLではCPUとGPUで同じプログラムを実行することでワークロードを共有できるため、プログラマーはデバイス間で作業を分割することで両方を活用できます。^[199]これにより、デバイス間で操作の相対速度が異なるため、作業をどのように分割するかを決定するという問題が生じます。この問題を解決するために機械学習が提案されています。GreweとO'Boyleは、プログラムのコンパイル時の特徴に基づいてトレーニングされたサポートベクターマシンのシステムを説明しています。このシステムは、実際にプログラムを実行してパフォーマンスを測定することなく、デバイスの分割問題を静的に決定できます。^[200]

AMD RDNA 2とNvidia RTXシリーズの実際のグラフィックカードの比較では、OpenCLテストによる結果は未確定です。Nvidia CUDAまたはOptiXの使用によるパフォーマンスの向上の可能性はテストされていません。^[201]