SYCL设备混合JIT与AOT编译：单设备AOT其余JIT可行吗？

可以实现，你遇到的PI_ERROR_INVALID_BINARY错误，是因为当前CMake配置将CPU的AOT二进制强制嵌入到了可执行文件中，导致GPU设备启动时尝试加载不兼容的CPU二进制，触发验证失败。以下是两种可行的解决思路：

方案一：分离AOT二进制与JIT编译逻辑 #

通过CMake单独编译CPU的AOT二进制文件，运行时根据设备类型动态选择加载AOT二进制或执行JIT编译：

1. CMake配置调整 #

# 主程序目标：通用JIT编译，适配所有设备
add_executable(sycl_app main.cpp)
target_compile_options(sycl_app PRIVATE -fsycl)
target_link_options(sycl_app PRIVATE -fsycl)

# 单独编译CPU AOT二进制文件
add_custom_command(
    OUTPUT cpu_aot.bin
    # 编译生成CPU目标的AOT对象文件
    COMMAND clang++ -fsycl -fsycl-targets=spir64_x86_64 -c main.cpp -o cpu_aot.o
    # 将对象文件提取为二进制资源
    COMMAND objcopy -O binary cpu_aot.o cpu_aot.bin
    DEPENDS main.cpp
)
add_custom_target(cpu_aot ALL DEPENDS cpu_aot.bin)

# 将AOT二进制复制到程序输出目录
add_custom_command(TARGET sycl_app POST_BUILD
    COMMAND ${CMAKE_COMMAND} -E copy ${CMAKE_CURRENT_BINARY_DIR}/cpu_aot.bin $<TARGET_FILE_DIR:sycl_app>
)

2. 代码中添加设备判断逻辑 #

#include <sycl/sycl.hpp>
#include <fstream>
#include <vector>

// 加载预编译的CPU AOT二进制
std::vector<char> load_cpu_aot_binary(const std::string& path) {
    std::ifstream file(path, std::ios::binary | std::ios::ate);
    std::streamsize size = file.tellg();
    file.seekg(0, std::ios::beg);
    std::vector<char> buffer(size);
    file.read(buffer.data(), size);
    return buffer;
}

int main() {
    sycl::device dev = sycl::default_selector_v.select_device();
    sycl::context ctx{dev};
    sycl::program prog{ctx};

    if (dev.is_cpu()) {
        // CPU设备加载AOT二进制
        auto cpu_binary = load_cpu_aot_binary("cpu_aot.bin");
        prog.build_from_binary(cpu_binary);
    } else {
        // GPU设备走JIT编译
        // 这里可以直接编译源代码，或从文件加载内核代码
        prog.build_with_source(R"(
            #include <sycl/sycl.hpp>

            template <int N>
            void kernel_func(sycl::handler& h) {
                h.parallel_for(sycl::range<1>(N), [=](sycl::id<1> idx) {
                    // 内核逻辑
                });
            }
        )");
    }

    // 创建并执行内核
    auto kernel = prog.get_kernel("kernel_func");
    sycl::queue q{ctx, dev};
    q.submit([&](sycl::handler& h) {
        h.use_kernel(kernel);
        // 绑定特化常量（仅JIT模式下生效）
        h.set_specialization_constant<0>(1024);
        kernel_func<1024>(h);
    });
    q.wait();

    return 0;
}

方案二：多目标混合编译（部分实现支持） #

如果使用Intel oneAPI等支持多目标编译的SYCL实现，可以通过编译选项指定仅对CPU做AOT，GPU保留JIT模式：

CMake配置示例 #

add_executable(sycl_app main.cpp)
target_compile_options(sycl_app PRIVATE 
    -fsycl 
    # 指定编译目标：CPU的SPIR-V和GPU的Gen架构IR
    -fsycl-targets=spir64_x86_64,spir64_gen 
    # 仅对CPU目标启用AOT编译
    -fsycl-aot-targets=spir64_x86_64
)
target_link_options(sycl_app PRIVATE 
    -fsycl 
    -fsycl-targets=spir64_x86_64,spir64_gen 
    -fsycl-aot-targets=spir64_x86_64
)

这种方式下，可执行文件会同时嵌入CPU的AOT二进制和GPU的中间表示（IR），SYCL runtime会自动根据设备类型选择对应的编译产物：CPU直接加载AOT二进制，GPU则对IR进行JIT编译，无需修改业务代码。

关键注意事项 #

• 特化常量仅在JIT模式下支持动态调整，CPU AOT编译时需提前确定特化常量的值（如果测试环境不需要动态修改，此限制不影响）。
• 不同SYCL厂商的AOT选项可能存在差异，例如Codeplay ComputeCpp使用-sycl-target和-sycl-aot参数，需参考对应厂商文档调整。
• 确保运行时设备检测逻辑准确，避免加载错误的二进制文件。

内容的提问来源于stack exchange，提问作者Fantastic Mr Fox