从 SPIR-V 到 ISPC:将 GPU 计算转化为 CPU 计算

游戏行业越来越多地趋向于将计算工作转移到图形处理单元 (GPU) 中,导致引擎和/或工作室需要开发大量 GPU 计算着色器来处理不同的计算任务。但有时候在 CPU 上运行这些计算着色器非常方便,不必重新投资开发它们的 C/C++ 变体。这样做的原因有很多,包括试验和调试非常简单,可充分利用备用 CPU 周期和鼓励基于 CPU 的内容扩展,与其他 CPU 端游戏资产之间进行基于 CPU 的交互,保证结果的确定性和一致性等等。

为帮助把握此次机会,同时充分利用内置于现代 CPU 内核中的单指令多数据 (SIMD) 矢量单元,我们基于开源 Khronos* SPIRV-Cross 项目,开始开发一款原型转换器,它将以标准可移植中间表示 (SPIR-V1) 为输入,输出 英特尔® SPMD 程序编译器(英特尔® SPC)(通常称为 ISPC)的内核。如下列使用示例所示,ispc.exe 采用 C 风格的内核,针对多种 ISA,比如英特尔® SIMD 流指令扩展(英特尔® SSE)、英特尔® 高级矢量扩展指令集 2(英特尔® AVX2)和英特尔® 高级矢量扩展 512(英特尔® AVX-512),生成高度矢量化的 CPU 对象文件。

该项目应视作转换的起点,而不是一款功能齐全,性能卓越的解决方案。目前,项目支持一部分标准 SPIR-V 内在函数、内置函数和类型,但它经过精心设计,可充分利用 ispc.exe 的内核性能特性,即统一性(标量)或多样性(矢量)变量。这有利于进行优化,比如在标量测试条件下避免出现多个昂贵的矢量分支。

代码在少量着色器上进行了测试,比如 Sascha Willems 的 Vulkan* 存储库中的计算示例和Microsoft DirectX* 12 MiniEngine 示例中的粒子系统计算着色器。

代码可从我们的 GitHub* 存储库下载,而且目前仅在 Windows* 系统上进行过测试。GitHub 自述文件包含更详细的文档,介绍了有关实施和支持特性的信息。

用法

glslangValidator.exe -H -V -o test.spv test.comp

spirv-cross.exe --ispc --output test.ispc test.spv

ispc.exe -O2 test.ispc -o test.ispc.obj -h test.ispc.h --target=avx2 --opt=fast-math

API 使用示例

ispc::raytracing_get_workgroup_size(workgroupSize[0], workgroupSize[1], workgroupSize[2]);

int32_t dispatch[3] = { textureComputeTarget.width / workgroupSize[0], textureComputeTarget.height / workgroupSize[1], 1 };
int32_t dispatch_count = dispatch[0] * dispatch[1];

concurrency::parallel_for<uint32_t>(0, dispatch_count, [&](uint32_t dispatchID)
{
    int32_t workgroupID[3] = { dispatchID % dispatch[0], dispatchID / dispatch[1], 0 };
    ispc::raytracing_dispatch_single(workgroupID, dispatch, planeCount, *pPlanes, sphereCount, *pSpheres, *pUBO, resultImage);
});

该项目在原始 SPIRV-Cross Apache 2.0 许可证下为开源项目,欢迎大家提供意见和建议。

如需进一步了解有关游戏中 ISPC 使用的信息,请阅读文章使用英特尔® SPMD 程序编译器在游戏中实现 CPU 矢量化

脚注

  1. SPIR-V 是面向 Vulkan 的默认着色器语言,可使用 glslangValidatorshaderc 编译器等工具,通过 OpenGL* 着色语言 (GLSL) 和高级着色语言 (HLSL) 着色器生成。
Para obtener información más completa sobre las optimizaciones del compilador, consulte nuestro Aviso de optimización.