Skip to content

Commit ebdf1c6

Browse files
authored
[clang][SYCL] Emit canonical types in integration footer (#16161)
A recent change to sycl headers revealed a bug in integration footer - missing nns in case of an explicit template specialization in some of device_global's template arguments.
1 parent 86376a8 commit ebdf1c6

File tree

2 files changed

+33
-0
lines changed

2 files changed

+33
-0
lines changed

clang/lib/Sema/SemaSYCL.cpp

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -6969,6 +6969,7 @@ bool SYCLIntegrationFooter::emit(raw_ostream &OS) {
69696969
Policy.adjustForCPlusPlusFwdDecl();
69706970
Policy.SuppressTypedefs = true;
69716971
Policy.SuppressUnwrittenScope = true;
6972+
Policy.PrintCanonicalTypes = true;
69726973

69736974
llvm::SmallSet<const VarDecl *, 8> Visited;
69746975
bool EmittedFirstSpecConstant = false;
Lines changed: 32 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,32 @@
1+
// RUN: %clang_cc1 -fsycl-is-device -internal-isystem %S/Inputs -triple spir64-unknown-unknown -fsycl-int-footer=%t.footer.h -emit-llvm %s -o -
2+
// RUN: FileCheck -input-file=%t.footer.h %s --check-prefix=CHECK-FOOTER
3+
4+
// This test checks that integration footer is emitted correctly when a
5+
// device_global has an explicit template specialization in template arguments.
6+
7+
#include "sycl.hpp"
8+
9+
namespace sycl {
10+
template <typename T> struct X {};
11+
template <> struct X<int> {};
12+
namespace detail {
13+
struct Y {};
14+
} // namespace detail
15+
template <> struct X<detail::Y> {};
16+
} // namespace sycl
17+
18+
using namespace sycl;
19+
template <typename T, typename = X<detail::Y>> struct Arg1 { T val; };
20+
21+
using namespace sycl::ext::oneapi;
22+
template <typename properties_t>
23+
device_global<properties_t> dev_global;
24+
25+
SYCL_EXTERNAL auto foo() {
26+
(void)dev_global<Arg1<int>>;
27+
}
28+
29+
// CHECK-FOOTER: __sycl_device_global_registration::__sycl_device_global_registration() noexcept {
30+
// CHECK-FOOTER-NEXT: device_global_map::add((void *)&::dev_global<Arg1<int, sycl::X<sycl::detail::Y>>>, "_Z10dev_globalI4Arg1IiN4sycl1XINS1_6detail1YEEEEE");
31+
// CHECK-FOOTER-NEXT: }
32+
// CHECK-FOOTER-NEXT: } // namespace (unnamed)

0 commit comments

Comments
 (0)