ワークアイテム・ワークグループ・次元数について

About

OpenCL では、次元数やワークグループ、ワークアイテムといった仕組みがあつかわれます(それぞれ dimention, work group, work item)。これらについては主に clEnqueueNDRangeKernel 関数で触れることになりますが、それぞれの関係が公式のリファレンスだけでは分かりにくいので、図を用意してまとめます。

ワークアイテム・ワークグループ

• ワークアイテム / work item
  • OpenCL では複数の処理をそれぞれのデバイス(プロセッサ)に並列して割り当てることで処理を最適化し高速化することを試みます。その分割の単位がワークアイテムです。つまりワークアイテムはカーネルです。

• ワークグループ / work group
  • 複数のワークアイテムを、いくつかのグループにしたものがワークグループです。

• LocalID・GrobalID
  • ワークアイテムにはそれぞれを識別するための ID が与えられます。ID はそれが含まれるワークグループ内で識別するための "LocalID" と、処理全体で識別するための "globalID" とがあります。

• local_work_size / work group size
  • ワークグループ内に含まれるワークアイテムの数は local_work_size(変数) もしくは work group size(表記上) で表されます。

• global_work_size
  • すべてのワークグループに存在するワークアイテムの数は global_work_size とされます。つまりワークアイテムの合計数です。

ここで clEnqueueNDRangeKernel 関数などを利用して、任意にワークアイテムを分割するとき、local_work_size は、CL_DEVICE_MAX_WORK_GROUP_SIZE で取得できる値よりも小さい値を指定する必要があります。また global_work_size を local_work_size で割り切れる必要があります。つまり、どのワークグループも同じ数のワークアイテムを持つ必要がある、ということです。

次元数(Dimention)

• work_dim / dimention
  • 先の図では、ワークアイテムとワークグループは横一列に並列されています。OpenCL ではN次元の並列化を指定することができます。並列化の次元数は work_dim で表されます。次元数は 0 より大きく、CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS で取得できる値以下である必要があります。ふつう 1～3 の値が入ります (OpenCL v1.2) 。

次元数を指定すると、ワークアイテムおよびワークグループが多次元化します。ID も合わせて多次元化する点に注意する必要があります。配列でイメージすると次のようになります。2次元の配列を指定すると、ワークグループは2次元になり、そのワークアイテムも2次元です。

• work_dim = 1
  • work_group[] = local_work_item[]
• work_dim = 2
  • work_group[,] = local_work_item[,]
• work_dim = 3
  • work_group[,,] = local_work_item[,,]