OpenCL 支持 ¶

Clang 完全支持 OpenCL C 1.0 到 3.0 版本。对 OpenCL 3.0 的支持处于实验阶段 (OpenCL 3.0).

有关在 OpenCL 中使用 clang 的详细信息，请参见 Clang 编译器用户手册.

缺少功能或支持有限 ¶

有关 clang 中 OpenCL 的一般问题和错误，请参阅 GitHub 问题列表.
命令行标志 -cl-ext（用于覆盖目标支持的扩展/功能）缺少对某些功能的支持，例如通过库完全实现的功能（参见基于库的功能和扩展）。

用假映射覆盖目标地址空间映射。这允许为默认情况下没有为每个 OpenCL 逻辑地址空间提供单独 ID 的非分段内存体系结构向位代码添加显式地址空间 ID。传递 -ffake-address-space-map 将使用以下值添加/覆盖为其编译的目标的地址空间：1-global、2-constant、3-local、4-generic。私有地址空间由 IR 中没有地址空间属性表示（另请参见有关地址空间属性的部分）。

$ clang -cc1 -ffake-address-space-map test.cl

OpenCL 内建函数 ¶

Clang 内建函数

有一些标准 OpenCL 函数作为 Clang 内建函数实现

OpenCL v2.0 内核语言规范第 6.13.16.2/6.13.16.3 节中的所有管道函数。
OpenCL v2.0 内核语言规范第 6.13.9 节中的地址空间限定符转换函数 to_global/to_local/to_private。
OpenCL v2.0 内核语言规范第 6.13.17.1 节中的所有 enqueue_kernel 函数以及 OpenCL v2.0 内核语言规范第 6.13.17.5 节中的查询函数。

快速内置函数声明

快速内置函数声明的实现（通过 -fdeclare-opencl-builtins 选项可用）包含以下主要组件

TableGen 定义文件 OpenCLBuiltins.td。它包含支持的内置函数的紧凑表示。添加新的内置函数声明时，通常只需要修改此文件。
在 ClangOpenCLBuiltinEmitter.cpp 中定义的 Clang TableGen 发射器。在 Clang 构建期间，发射器读取 TableGen 定义文件并生成 OpenCLBuiltins.inc。此生成的文件包含各种表和函数，这些表和函数以紧凑的方式捕获 TableGen 定义中的内置函数数据。
SemaLookup.cpp 中的 OpenCL 特定代码。当 Sema::LookupBuiltin 遇到潜在的内置函数时，它将检查该名称是否对应于有效的 OpenCL 内建函数。如果是，则使用 InsertOCLBuiltinDeclarationsFromTable 插入该函数的所有重载，然后进行重载解析。

OpenCL 扩展和功能 ¶

Clang 实现 OpenCL 内核语言的各种扩展。

只要文档的详细程度足以实施，新功能就会被接受。应该有证据表明，该扩展是考虑到实施可行性而设计的，并且对基于 C/C++ 的编译器的复杂性进行了评估。或者，可以接受草案形式的文档，可以在实施过程中对其进行进一步完善。

实施指南 ¶

本节说明如何使用新功能扩展 clang。

解析功能

如果扩展修改了标准解析，则需要将其添加到 clang 前端源代码中。这也意味着应该将指示扩展存在的相关宏添加到 clang 中。

将新扩展添加到前端的默认流程是修改 OpenCLExtensions.def，其中包含前端支持的所有扩展和可选功能的列表。

这将自动添加宏，并在目标选项 clang::TargetOptions::OpenCLFeaturesMap 中添加一个字段，以控制编译期间新扩展的公开方式。

请注意，默认情况下，SPIR-V、SPIR 或 X86 等目标会公开所有 OpenCL 扩展。对于所有其他目标，必须明确进行配置。

请注意，clang 执行的目标扩展支持可以使用 -cl-ext 命令行标志覆盖。

库功能

如果扩展添加的功能不修改标准语言解析，则不需要修改除头文件和 OpenCLBuiltins.td 以外的任何内容，详细说明见 OpenCL 内置函数。通常，此类扩展通过库添加功能（通过添加非原生类型或函数），这些功能被定期解析。类似于其他语言，这是添加新功能的最常见方法。

Clang 有标准头文件，在其中添加了新的类型和函数，有关更多详细信息，请参考 OpenCL 头文件部分。指示此类扩展存在情况的宏可以添加到标准头文件中，这些宏以目标特定的预定义宏或/和语言版本预定义宏为条件（请参阅在 opencl-c-base.h 中定义的 feature/extension 预处理器宏）。

编译指示

一些扩展通过编译指示动态地修改标准解析。

Clang 提供了一种机制来添加标准扩展编译指示 OPENCL EXTENSION，方法是在 OpenCLExtensions.def 的扩展列表条目中设置一个专用标志。请注意，标准扩展编译指示没有默认行为，因为它在标准规范（包括版本 3.0 在内）中没有充分详细地说明，因此 clang 没有提供默认功能。

没有详细行为信息的编译指示（例如，它触发解析更改的解释）不应该添加到 clang 中。此外，编译指示应该为用户提供有用的功能。例如，此类功能应该解决实际用例，而不是冗余，即不能使用现有功能实现。

请注意，一些遗留扩展（在 OpenCL 3.0 之前发布）仍然为编译指示提供了一些不符合标准的功能，例如，添加有关类型或函数使用的诊断信息。此功能不保证在将来的版本中保留。但是，任何未来的更改不应影响向后兼容性。

地址空间属性 ¶

Clang 使用 address_space(N) 属性支持任意地址空间，其中 N 是 Clang 源代码中指定的范围内的整数。此地址空间可以与 OpenCL 地址空间一起使用，但是当此类地址空间转换为/从 OpenCL 地址空间转换时，行为不受 OpenCL 规范的约束。

OpenCL 实现使用关键字提供标准地址空间列表：private、local、global 和 generic。在 AST 和 IR 中，每个地址空间将由 Clang 源代码中提供的唯一编号表示。地址空间的特定 ID 不必在 AST 和 IR 之间匹配。通常，在 AST 中，地址空间编号代表逻辑段，而在 IR 中，它们代表物理段。因此，具有扁平内存段的机器可以将所有 AST 地址空间编号映射到同一个物理段 ID 或在生成 IR 时完全跳过地址空间属性。但是，如果 IR 传递需要地址空间信息，例如，为了改进别名分析，建议保留它，并且只降低它以在晚期的机器传递中反映物理内存段。逻辑地址空间和目标地址空间之间的映射在 Clang 的源代码中指定。

面向 OpenCL 的 C++ 实现状态 ¶

Clang 实现发布在面向 OpenCL 的 C++ 文档的官方版本中的语言版本 1.0 和 2021。

实验性特性中描述了对实验性 C++ 库的有限支持。

此功能的 GitHub 问题通常以“[C++4OpenCL]”为前缀 - 点击这里查看完整的错误列表。

缺少的特性或支持有限的特性 ¶

目前，对面向 OpenCL 的 C++ 2021 的支持处于实验阶段。有关 OpenCL 3.0 中常见的缺失功能的详细信息，请参阅 OpenCL 3.0 状态。
针对非平凡全局析构函数的 IR 生成是不完整的（请参阅：PR48047）。
对具有非默认地址空间的析构函数的支持是不完整的（请参阅：D109609）。

OpenCL C 3.0 使用 ¶

OpenCL C 3.0 语言标准使大多数 OpenCL C 2.0 功能可选。OpenCL C 3.0 中的可选功能通过特征测试宏（特征测试宏列表在这里）来指示。命令行标志 -cl-ext 可用于覆盖目标支持的特性。

对于存在特定特性相关扩展的情况（fp64 和 3D 图像写入），用户应在命令行标志中同时指定两者（扩展和特性）

$ clang -cl-std=CL3.0 -cl-ext=+cl_khr_fp64,+__opencl_c_fp64 ...
$ clang -cl-std=CL3.0 -cl-ext=-cl_khr_fp64,-__opencl_c_fp64 ...

OpenCL C 3.0 实现状态 ¶

下表概述了 OpenCL C 3.0 中的特性及其实现状态。

类别	特性	状态	评审
命令行界面	`-cl-std` 标志的新值	完成	https://reviews.llvm.org/D88300
预定义宏	新的版本宏	完成	https://reviews.llvm.org/D88300
预定义宏	特性宏	完成	https://reviews.llvm.org/D95776
特性可选性	通用地址空间	完成	https://reviews.llvm.org/D95778 和 https://reviews.llvm.org/D103401
特性可选性	带有通用地址空间的内置函数重载	完成	https://reviews.llvm.org/D105526、https://reviews.llvm.org/D107769
特性可选性	全局内存中的程序范围变量	完成	https://reviews.llvm.org/D103191
特性可选性	包括内置函数的 3D 图像写入	完成	https://reviews.llvm.org/D106260（前端）
特性可选性	包括内置函数的读写图像	完成	https://reviews.llvm.org/D104915（前端）和 https://reviews.llvm.org/D107539、https://reviews.llvm.org/D117899（函数）
特性可选性	C11 原子内存范围、排序和内置函数	完成	https://reviews.llvm.org/D106111、https://reviews.llvm.org/D119420
特性可选性	包括内置函数的块和设备端内核入队	完成	https://reviews.llvm.org/D115640、https://reviews.llvm.org/D118605
特性可选性	包括内置函数的管道	完成	https://reviews.llvm.org/D107154（前端）和 https://reviews.llvm.org/D105858（函数）
特性可选性	工作组集体内置函数	完成	https://reviews.llvm.org/D105858
特性可选性	图像类型和内置函数	完成	https://reviews.llvm.org/D103911（前端）和 https://reviews.llvm.org/D107539（函数）
特性可选性	双精度浮点类型	完成	https://reviews.llvm.org/D96524
新功能	RGBA 向量分量	完成	https://reviews.llvm.org/D99969
新功能	子组函数	完成	https://reviews.llvm.org/D105858、https://reviews.llvm.org/D118999
新功能	原子内存范围：子组、所有设备，包括函数	完成	https://reviews.llvm.org/D103241

实验性特性 ¶

Clang 为开发者提供了以下新的 WIP 特性，以便他们进行实验并提供早期反馈或通过进一步改进做出贡献。随时通过 Discourse 论坛（Clang 前端类别）联系我们，或提交 GitHub 问题。

面向 OpenCL 的 C++ 库 ¶

目前正在进行一项工作，以使用面向 OpenCL 的 C++ 模式，在 OpenCL 内核代码中支持来自 LLVM 的 libcxx 的 C++ 标准库。

当以下 clang 扩展启用时，目前可以在内核源代码中包含 C++17 中的 type_traits，__cl_clang_function_pointers 和 __cl_clang_variadic_functions，有关更多详细信息，请参阅 Clang 语言扩展。由扩展启用的不符合标准特性的使用不会暴露超出编译范围的不符合标准行为，即不会在 IR 或二进制文件中生成。该扩展仅出现在元编程机制中，用于识别或验证类型的属性。这允许在不损失可移植性的情况下提供完整的 C++ 功能。为了避免不安全地使用扩展，建议在包含头文件后立即禁用扩展。

使用示例:

以下说明了包含 type_traits 的内核代码示例。

#pragma OPENCL EXTENSION __cl_clang_function_pointers : enable
#pragma OPENCL EXTENSION __cl_clang_variadic_functions : enable
#include <type_traits>
#pragma OPENCL EXTENSION __cl_clang_function_pointers : disable
#pragma OPENCL EXTENSION __cl_clang_variadic_functions : disable

using sint_type = std::make_signed<unsigned int>::type;

__kernel void foo() {
  static_assert(!std::is_same<sint_type, unsigned int>::value);
}

编译示例的可能 clang 调用如下所示

$ clang -I<path to libcxx checkout or installation>/include test.clcpp

请注意，type_traits 是一个仅包含头文件的库，因此不需要针对标准库进行额外的链接步骤。在 Compiler Explorer 中查看完整的示例。

更多基于 libcxx 构建的 OpenCL 特定 C++ 库实现可在 libclcxx 项目中找到。

Clang 20.0.0git 文档

OpenCL 支持