カーネルは以下のように、__kernelという修飾子が付き、さらに戻り値型は常にvoid型の関数を指します。以下がカーネル関数の例です。

__kernel void func(args) { //(1)

}

前の例では引数がargsとなっていますが、C言語と同様にOpenCL-Cでは、型名と変数名を引数で指定します。

__kernel void func(__global float* data) { //(1)

}

この例ではグローバルメモリを指定しているので、引数に渡されたdataはメモリオブジェクトになります。メモリオブジェクトはカーネル内ではなく、カーネルを呼び出すホストアプリケーションで生成し、これをカーネルの引数として複製・転送します。

データ並列プログラミングやSPMDの概念では、同一カーネルを実行すると解説してきましたが、同時にカーネルは固有のカウンタを持つという定義がありました。

これをカーネル関数で実現する場合には、まずは以下の関数を使ってカーネルのカウンタであるグローバルIDを取得します。

size_t get_global_id(uint dimindx)

引数のdimindxは次元１の場合は常に「0」の値となります。2次元の場合は、「0」か「1」のいずれかを指定して、取得したいグローバルIDがどの次元に属するかを設定します。関数が戻す値は、ワークアイテムに固有に割り振られる整数型（size_t）のグローバルIDです。

以下のカーネル関数では、グローバルIDをidという変数に代入して、idをグローバルメモリ領域に確保した配列の要素インデックス（添字）として使います。

__kernel void func(__global float* data) {
    size_t id = get_global_id(0); //(1)
    data[id] = data[id] * data[id]; //(2)
}