clang - spir-v バイナリカーネルで clCreateProgramWithIL() を適切に使用するには?

Question

いくつかのエキゾチックなプロファイリングの目的で、入力として受け取る OpenCL カーネルのコードを計測する OpenCL アプリケーションを作成する必要があります (必要なものが見つからないため、自分で行う必要がある/したい)。

カーネルを中間表現 (現在は LLVM-IR) にコンパイルし、インストルメント化 (LLVM C++ バインディングを使用) し、インストルメント化されたコードを SPIR-V にトランスパイルしてから、ホストコードでclCreateProgramWithIL().

今のところ、インストルメンテーションなしで 2 つのベクトルを追加する単純な OpenCL カーネルをコンパイルしています。

__kernel void vadd(
    __global float* a,
    __global float* b,
    __global float* c,
    const unsigned int count)
{
    int i = get_global_id(0);
    if(i < count) c[i] = a[i] + b[i];
}

上記を LLVM IR にコンパイルするには、次のコマンドを使用します。

clang -c -emit-llvm -include libclc/generic/include/clc/clc.h -I libclc/generic/include/ vadd.cl -o vadd.bc -emit-llvm -O0 -x cl

その後、ツール( herevadd.bc )を使用してトランスパイルします。vadd.spvllvm-spirv

最後に、次のように C ホストコードからカーネルを構築してみます。

...
cl_program program = clCreateProgramWithIL(context, binary_data->data, binary_data->size, &err);
err = clBuildProgram(program, 1, &device_id, NULL, NULL, NULL);
...

clBuildProgramホストコードを実行した後、コマンドから上記のエラーが表示されます。

CL_BUILD_PROGRAM_FAILURE
error: undefined reference to `get_global_id()'
error: backend compiler failed build.

vadd.spvファイルが OpenCL カーネル ライブラリとリンクされていないようです。これを達成する方法はありますか？