\section{OpenCL}
OpenCL とは、Multi Core CPU と GPU のようなヘテロジニアスな環境を利用した並列計算を支援するフレームワークである。
演算用のプロセッサ(本研究では GPU)上で動作するプログラム OpenCL C という言語で記述する。
OpenCL C で記述したプログラムを GPU 上で実行させるために OpenCL Runtime API を利用する。
OpenCL ではオペレーティングシステムなどが処理されるメイン CPU のことを host、GPGPU が可能なグラフィックボードなどのことを device と定義している。
OpenCL Application は host 側のプログラムと device 側のプログラムが一体となって動作する。
この device 上で動作するプログラムを特別に kernel と呼ぶ。

\subsection{CommandQueue}
OpenCL では、device の操作に CommandQueue を使用する。
CommandQueue は device に Operation を送るための仕組みである。
kernel の実行、input buffer の読み込み、output buffer への書き込みなどが Operation となる。

CommandQueue に投入された Operation は投入された順序で実行される。
CommandQueue を生成するときプロパティを指定することで Operation を投入された順序を無視して(out of order)実行することが可能になる。
また複数の CommandQueue を生成し、device に投入することでも our of order で実行することが可能である。

out of order で実行する場合、データの依存関係を設定する必要がある。各 Operation を発行する関数には event\_wait\_list と event を指定することができ、これらを利用してデータの依存関係を設定することができる。
out of order 実行を可能にするプロパティをサポートしている device が少ないため、今回は複数の CommandQueue を用いる方法で実装を行なった。

\subsection{Data Parallel Execution}
3D グラフィックのような多次元のデータを処理する場合に高い並列度を保つには、データを分割して並列に実行する機能が必要である。
これを OpenCL ではデータ並列と呼んでいる。
OpenCL では次元数に対応する index があり、OpenCL は1つの記述から index の異なる複数の kernel を自動生成する。
その添字を global\_id と呼ぶ。
このとき入力されたデータは WorkItem という処理単位に分割される。

OpenCL は WorkItem に対して、それぞれを識別する ID(global\_id)を割り当てる。
kernel は get\_global\_id という API によって ID を取得し、取得した ID に対応するデータに対して処理を行うことでデータ並列を実現する。

データ並列による kernel 実行の場合、clEnqueueNDRangeKernel を使用する。
この関数の引数として WorkItem の数と次元数を指定することでデータ並列で実行することができる。