読者です 読者をやめる 読者になる 読者になる

溜息日和

プログラミングとか機械学習とか統計力学とかいろいろ

OpenCLの言葉

NDRange

ワークグループをまとめたインデックス空間のこと

オブジェクト

Kernel

データに施す演算・処理の単位

Program

カーネルの集合・デバイス・リスト・実行バイナリ・ソースの集合

Context

各デバイスのカーネル・メモリ処理を定義するもの
バイスに対するカーネルの実行環境を定義したものである.
ホストはOpenCLバイス毎にコンテキストを定義し,コンテキストに対して

  • カーネルオブジェクト
  • メモリオブジェクト
  • プログラムオブジェクト

などのカーネルの実行に必要なオブジェクトを作成する.

Queue

様々な処理の実行を司る仕組み(Contextに届く指令書)

Buffer

カーネルがアクセスする一定量のメモリのかたまり

管理階層

各層が多次元で定義することが可能であり,それぞれに多次元のIDが割り振られている.
このIDを取得する関数の引数には,取得したいIDの次元を代入すれば良い.

// デバイスの情報を取得するAPI
clGetDeviceInfo (device, param_name, param_value_size, *param_value, *param_value_size_ret)

param_name = CL_DEVICE_MAX_COMPUTE_UNITSとすることで,並列演算のコア数がわかる.
param_name = CL_DEVICE_MAX_WORK_ITEM_DIMENSIONとすることで,ワークアイテムの次元の上限がわかる.
param_name = CL_DEVICE_MAX_WORK_ITEM_SIZEとすることで,ワークグループ毎のワークアイテム数の上限がわかる.

Work Group

Work Itemを集合
演算ユニットで実行
ワークグループあたりの,ワークアイテム数は,全てのワークグループで同じになる
全体のワークアイテム数が,ワークグループあたりのワークアイテム数で,割り切れない場合,clNDEnqueueNDRangeKernelは失敗する.

get_group_id() // ワークグループIDを取得する
Work Item

処理データの最小単位
プロセッシングエレメントで実行

get_local_id()   // あるWork Groupの中のWork ItemのIDを取得する 
Global Work Item

Work Groupの中のWork Itemと捉えずに,Work Groupに囚われない特定のGlobal Work Item

get_global_id() // 特定のWork ItemのIDを取得する 

ベクタデータ

SIMD計算に自動的に割り振られる
SIMD演算器を持つか否かは,clGetDeviceInfoで取ってくる
型名の後に数字(2, 3, 4, 8, 16)をつければベクタ型になる.

int4    hoge4; 
char2  piyo2;
float8 foo8;

ベクタリテラル構文を用いて,ベクタ値を使う
演算・関数は各要素毎に適用される

int4 hoge4 = (int4)(1.0f, 2.0f, 3.0f, 4.0f);
int4 piyo4  = log(hoge4);
ベクタデータの参照
  • 変数.s参照したい要素番号0〜f
  • 変数.xyzw(4番目以降の要素は参照できない)
  • 変数.(odd:奇数 or even:偶数 or hi:上半分 or lo:下半分)
int4 piyo4 = (int4)(1.0f, 2.0f, 3.0f, 4.0f);
int3 hoge4 = piyo4.302;
int4 hoge4 = piyo4.xyzw;
int2 hoge2 = piyo4.even;
ベクタデータの比較演算
  • オペランドの要素の型が変形されることがある.
  • 各要素毎に比較を行いt = -1, f = 0を返す

ホストプログラム

制御用プログラム
カーネルプログラムの実行制御のために様々な手続きを行う
ホストプログラムAPIの雑なまとめ

定義順 定義名 API
1 OpenCLが起動するプラットフォーム(Nvidia, AMD, Intel, Apple)の特定する clGetPlatformIDs(*, &platform_id, *)
2 プラットフォームを構成するハードウェアの特定 clGetDeviceIDs(platform_id, *, *, &device_id, *)
3 OpenCLを実行させる環境を構築 context = clCreateContext(*, *, &device_id, *, *, *)
4 バイスを制御するコマンドキューを作成 command_queue = clCreateCommandQueue(context, device_id, *, *)
5 バイス側メモリをホスト側で認識 memobj = clCreateBuffer(context, *, *, *, *)
6 カーネルコードをカーネル関数ごとに整理しプログラムオブジェクトの作成 program = clCreateProgramWithSource(context, *, *, *, *)
7 プログラムオブジェクトをバイナリに変換 clBuildProgram(program, *, &device_id, *, *, *)
8 プログラムオブジェクトをカーネル関数と連関 kernel = clCreateKernel(program, *, *)
9 カーネル関数の引数の設定 clSetKernelArg(kernel, *, *, (void *)&memobj)
10 カーネルの実行 clEnqueueTask(command_queue, kernel, *, *, *)
11 カーネルの実行により作成されたデバイス側メモリをホスト側メモリに転送する clEnqueueReadBuffer(command_queue, memobj, *, *, *, *, *, *, *)
12 各オブジェクトの開放 clRelease***(***)

メモリ階層

メモリ内容の一貫性はその管理階層内でしか保証されない.よって,Global - Constant - Local - Privateの各階層間のデータ転送には,明確な配慮が必要である.

Global Memory

アドレス空間修飾子: __global
ホスト・カーネルからの読み書き可

Constant Memory

アドレス空間修飾子: __constant
ホストからの読み書き込み可
カーネルからの書き込み不可
カーネルの実行中に定数を保持するための領域
ホストプログラムはメモリオブジェクトをアロケート,初期化した後に,Constant Memoryに配置する.

Local Memory

アドレス空間修飾子: __local
ホストからアクセス不可
カーネルから読み書き可
Work Group内で共有される

Private Memory

アドレス空間修飾子: private
ホストからアクセス不可
カーネルからの読み書き可
Work Item単位のメモリ

広告を非表示にする