[Clang][OpenCL] Fix Missing -fdeclare-opencl-builtins When Using --save-temps

[shiltian]

When compiling an OpenCL program directly with clang using --save-temps, an
error may occur if the program contains OpenCL builtins:

test.cl:3:21: error: use of undeclared identifier 'get_global_id'
    3 |   unsigned int id = get_global_id(0);
      |                     ^

This happens because the driver does not add -fdeclare-opencl-builtins when
the input type is TY_PP_CL. This PR fixes the issue.

[shiltian]

• [Clang][OpenCL] Fix Missing -fdeclare-opencl-builtins When Using --save-temps #131017 👈 (View in Graphite)
• main

This stack of pull requests is managed by Graphite. Learn more about stacking.

[llvmbot]

@llvm/pr-subscribers-clang-driver

@llvm/pr-subscribers-clang

Author: Shilei Tian (shiltian)

Changes

When compiling an OpenCL program directly with clang using --save-temps, an
error may occur if the program contains OpenCL builtins:

test.cl:3:21: error: use of undeclared identifier 'get_global_id'
    3 |   unsigned int id = get_global_id(0);
      |                     ^

This happens because the driver does not add -fdeclare-opencl-builtins when
the input type is TY_PP_CL. This PR fixes the issue.

---

Full diff: https://github.com/llvm/llvm-project/pull/131017.diff

2 Files Affected:

• (modified) clang/lib/Driver/Types.cpp (+11-1)
• (added) clang/test/Driver/opencl-aot-compilation.cl (+4)

diff --git a/clang/lib/Driver/Types.cpp b/clang/lib/Driver/Types.cpp
index eaca74a7b5529..08866e89ea447 100644
--- a/clang/lib/Driver/Types.cpp
+++ b/clang/lib/Driver/Types.cpp
@@ -226,7 +226,17 @@ bool types::isObjC(ID Id) {
   }
 }
 
-bool types::isOpenCL(ID Id) { return Id == TY_CL || Id == TY_CLCXX; }
+bool types::isOpenCL(ID Id) {
+  switch (Id) {
+  default:
+    return false;
+  case TY_PP_CL:
+  case TY_PP_CLCXX:
+  case TY_CL:
+  case TY_CLCXX:
+    return true;
+  }
+}
 
 bool types::isCXX(ID Id) {
   switch (Id) {
diff --git a/clang/test/Driver/opencl-aot-compilation.cl b/clang/test/Driver/opencl-aot-compilation.cl
new file mode 100644
index 0000000000000..eba658619b7f8
--- /dev/null
+++ b/clang/test/Driver/opencl-aot-compilation.cl
@@ -0,0 +1,4 @@
+// RUN: %clang -x cl %s --save-temps -### 2>&1 | FileCheck %s
+
+// CHECK: "-fdeclare-opencl-builtins" {{.*}} "[[SRC:.+]].cli" "-x" "cl" "{{.*}}[[SRC]].cl"
+// CHECK-NEXT: "-fdeclare-opencl-builtins" {{.*}} "-x" "cl-cpp-output" "[[SRC]].cli"

[jhuber6]

Makes sense, though I don't know why we have a special flag for what is essentially a header.

[shiltian]

Yeah these are implemented in bitcode file, therefore it needs the front end to be able to recognize it instead of treating it as an unknown symbol.

[jhuber6]

Yeah these are implemented in bitcode file, therefore it needs the front end to be able to recognize it instead of treating it as an unknown symbol.

In a normal world they would just be in a header and then the library would get linked in later.

[arsenm]

These are not implemented in Bitcode linked in later. Tablegen emits a table clang directly consumes to pre-declare since it's faster than parsing the builtin header