From a6f5b4e00a1f688081a9a9d49ce0940403901a87 Mon Sep 17 00:00:00 2001 From: Dmitry Vodopyanov Date: Tue, 18 May 2021 18:29:58 +0300 Subject: [PATCH 01/21] [SYCL] Move __buildin_unique_stable_name dependent funcs to a separate header This patch moves functions which use `__builtin_unique_stable_name` to a separate header which should only be included to integration footer. This was made to get rid of UB when the compiler sees defenition of the template specialization before seeing a declaration. --- sycl/include/CL/sycl/detail/kernel_desc.hpp | 13 +++---- .../CL/sycl/detail/spec_const_integration.hpp | 35 +++++++++++++++++++ 2 files changed, 39 insertions(+), 9 deletions(-) create mode 100644 sycl/include/CL/sycl/detail/spec_const_integration.hpp diff --git a/sycl/include/CL/sycl/detail/kernel_desc.hpp b/sycl/include/CL/sycl/detail/kernel_desc.hpp index 7db5b75386e17..1d6fbd46d2881 100644 --- a/sycl/include/CL/sycl/detail/kernel_desc.hpp +++ b/sycl/include/CL/sycl/detail/kernel_desc.hpp @@ -21,7 +21,8 @@ namespace detail { // while the code requires C++17. This code is not supposed to be used by the // libsycl.so so it should not be a problem. #if __cplusplus > 201402L -template struct specialization_id_name_generator {}; +// Definition is in spec_const_integration.hpp +template struct specialization_id_name_generator; #endif #ifndef __SYCL_DEVICE_ONLY__ @@ -58,14 +59,8 @@ template struct SpecConstantInfo { #if __cplusplus >= 201703L // Translates SYCL 2020 specialization constant type to its name. -template const char *get_spec_constant_symbolic_ID() { -#ifdef SYCL_LANGUAGE_VERSION - return __builtin_unique_stable_name( - specialization_id_name_generator); -#else - return ""; -#endif -} +// Definition is in spec_const_integration.hpp +template const char *get_spec_constant_symbolic_ID(); #endif #ifndef __SYCL_UNNAMED_LAMBDA__ diff --git a/sycl/include/CL/sycl/detail/spec_const_integration.hpp b/sycl/include/CL/sycl/detail/spec_const_integration.hpp new file mode 100644 index 0000000000000..7a9d0354bf233 --- /dev/null +++ b/sycl/include/CL/sycl/detail/spec_const_integration.hpp @@ -0,0 +1,35 @@ +//==---------------------- spec_const_integration.hpp ----------------------==// +// +// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. +// See https://llvm.org/LICENSE.txt for license information. +// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception +// +//===----------------------------------------------------------------------===// + +#pragma once + +// This header file must not be included to any DPC++ headers. +// This header file should only be included to integration footer. + +__SYCL_INLINE_NAMESPACE(cl) { +namespace sycl { +namespace detail { + +#if __cplusplus >= 201703L +template struct specialization_id_name_generator {}; +#endif + +// Translates SYCL 2020 specialization constant type to its name. +template const char *get_spec_constant_symbolic_ID() { +#ifdef SYCL_LANGUAGE_VERSION + return __builtin_unique_stable_name( + specialization_id_name_generator); +#else + return ""; +#endif +} +#endif + +} // namespace detail +} // namespace sycl +} // __SYCL_INLINE_NAMESPACE(cl) \ No newline at end of file From dd66b4434d05508509d79f75278f6c7cb5631e2b Mon Sep 17 00:00:00 2001 From: Dmitry Vodopyanov Date: Tue, 18 May 2021 18:50:38 +0300 Subject: [PATCH 02/21] Add missing EOL --- sycl/include/CL/sycl/detail/spec_const_integration.hpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/sycl/include/CL/sycl/detail/spec_const_integration.hpp b/sycl/include/CL/sycl/detail/spec_const_integration.hpp index 7a9d0354bf233..5012de72dd3ba 100644 --- a/sycl/include/CL/sycl/detail/spec_const_integration.hpp +++ b/sycl/include/CL/sycl/detail/spec_const_integration.hpp @@ -32,4 +32,4 @@ template const char *get_spec_constant_symbolic_ID() { } // namespace detail } // namespace sycl -} // __SYCL_INLINE_NAMESPACE(cl) \ No newline at end of file +} // __SYCL_INLINE_NAMESPACE(cl) From 22f900922a49d40f53b0f57c8e0bc13b400d98ba Mon Sep 17 00:00:00 2001 From: Dmitry Vodopyanov Date: Tue, 18 May 2021 18:52:50 +0300 Subject: [PATCH 03/21] Remove extra endif --- sycl/include/CL/sycl/detail/spec_const_integration.hpp | 1 - 1 file changed, 1 deletion(-) diff --git a/sycl/include/CL/sycl/detail/spec_const_integration.hpp b/sycl/include/CL/sycl/detail/spec_const_integration.hpp index 5012de72dd3ba..e25fe8c549bae 100644 --- a/sycl/include/CL/sycl/detail/spec_const_integration.hpp +++ b/sycl/include/CL/sycl/detail/spec_const_integration.hpp @@ -17,7 +17,6 @@ namespace detail { #if __cplusplus >= 201703L template struct specialization_id_name_generator {}; -#endif // Translates SYCL 2020 specialization constant type to its name. template const char *get_spec_constant_symbolic_ID() { From 1bd1e947e105a84aff39a73450bed6594d7ed7ca Mon Sep 17 00:00:00 2001 From: Dmitry Vodopyanov Date: Mon, 31 May 2021 16:12:34 +0300 Subject: [PATCH 04/21] Add -fsycl-use-footer to tests and include missing header to int footer --- clang/lib/Sema/SemaSYCL.cpp | 1 + .../specialization_constants/non_native/accelerator.cpp | 2 +- .../basic_tests/specialization_constants/non_native/cpu.cpp | 2 +- .../basic_tests/specialization_constants/non_native/cuda.cpp | 2 +- .../basic_tests/specialization_constants/non_native/gpu.cpp | 2 +- 5 files changed, 5 insertions(+), 4 deletions(-) diff --git a/clang/lib/Sema/SemaSYCL.cpp b/clang/lib/Sema/SemaSYCL.cpp index c2c11765a8bc2..139b0a1db08c6 100644 --- a/clang/lib/Sema/SemaSYCL.cpp +++ b/clang/lib/Sema/SemaSYCL.cpp @@ -4897,6 +4897,7 @@ bool SYCLIntegrationFooter::emit(raw_ostream &OS) { for (const VarDecl *VD : SpecConstants) { VD = VD->getCanonicalDecl(); std::string TopShim = EmitSpecIdShims(OS, ShimCounter, VD); + OS << "#include \n"; OS << "__SYCL_INLINE_NAMESPACE(cl) {\n"; OS << "namespace sycl {\n"; OS << "namespace detail {\n"; diff --git a/sycl/test/on-device/basic_tests/specialization_constants/non_native/accelerator.cpp b/sycl/test/on-device/basic_tests/specialization_constants/non_native/accelerator.cpp index c11958d10d07a..081b7c3242cea 100644 --- a/sycl/test/on-device/basic_tests/specialization_constants/non_native/accelerator.cpp +++ b/sycl/test/on-device/basic_tests/specialization_constants/non_native/accelerator.cpp @@ -1,6 +1,6 @@ // REQUIRES: aoc, accelerator -// RUN: %clangxx -fsycl -fsycl-targets=spir64_fpga-unknown-unknown-sycldevice %S/Inputs/common.cpp -o %t.out +// RUN: %clangxx -fsycl -fsycl-use-footer -fsycl-targets=spir64_fpga-unknown-unknown-sycldevice %S/Inputs/common.cpp -o %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out // This test checks correctness of SYCL2020 non-native specialization constants diff --git a/sycl/test/on-device/basic_tests/specialization_constants/non_native/cpu.cpp b/sycl/test/on-device/basic_tests/specialization_constants/non_native/cpu.cpp index a2c91f1b58de3..555f7801bcd75 100644 --- a/sycl/test/on-device/basic_tests/specialization_constants/non_native/cpu.cpp +++ b/sycl/test/on-device/basic_tests/specialization_constants/non_native/cpu.cpp @@ -1,6 +1,6 @@ // REQUIRES: opencl-aot, cpu -// RUN: %clangxx -fsycl -fsycl-targets=spir64_x86_64-unknown-unknown-sycldevice %S/Inputs/common.cpp -o %t.out +// RUN: %clangxx -fsycl -fsycl-use-footer -fsycl-targets=spir64_x86_64-unknown-unknown-sycldevice %S/Inputs/common.cpp -o %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out // This test checks correctness of SYCL2020 non-native specialization constants diff --git a/sycl/test/on-device/basic_tests/specialization_constants/non_native/cuda.cpp b/sycl/test/on-device/basic_tests/specialization_constants/non_native/cuda.cpp index 01d0d37e04f26..5229c3c2d5fc4 100644 --- a/sycl/test/on-device/basic_tests/specialization_constants/non_native/cuda.cpp +++ b/sycl/test/on-device/basic_tests/specialization_constants/non_native/cuda.cpp @@ -1,6 +1,6 @@ // REQUIRES: cuda -// RUN: %clangxx -fsycl -fsycl-targets=nvptx64-unknown-unknown-sycldevice %S/Inputs/common.cpp -o %t.out +// RUN: %clangxx -fsycl -fsycl-use-footer -fsycl-targets=nvptx64-unknown-unknown-sycldevice %S/Inputs/common.cpp -o %t.out // RUN: env SYCL_DEVICE_FILTER=cuda %t.out // TODO: enable this test then compile-time error in sycl-post-link is fixed diff --git a/sycl/test/on-device/basic_tests/specialization_constants/non_native/gpu.cpp b/sycl/test/on-device/basic_tests/specialization_constants/non_native/gpu.cpp index df9f0a7937dc3..84e56e51bc7a5 100644 --- a/sycl/test/on-device/basic_tests/specialization_constants/non_native/gpu.cpp +++ b/sycl/test/on-device/basic_tests/specialization_constants/non_native/gpu.cpp @@ -2,7 +2,7 @@ // UNSUPPORTED: cuda // CUDA is not compatible with SPIR. -// RUN: %clangxx -fsycl -fsycl-targets=spir64_gen-unknown-unknown-sycldevice -Xsycl-target-backend=spir64_gen-unknown-unknown-sycldevice "-device *" %S/Inputs/common.cpp -o %t.out +// RUN: %clangxx -fsycl -fsycl-use-footer -fsycl-targets=spir64_gen-unknown-unknown-sycldevice -Xsycl-target-backend=spir64_gen-unknown-unknown-sycldevice "-device *" %S/Inputs/common.cpp -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // This test checks correctness of SYCL2020 non-native specialization constants From 1bd83e012d8f54e1463566992b18ebf845ca478a Mon Sep 17 00:00:00 2001 From: Dmitry Vodopyanov Date: Tue, 1 Jun 2021 20:31:44 +0300 Subject: [PATCH 05/21] Enable __builtin_unique_stable_name in FE; Make -fsycl-use-footer enabled by default for any app --- clang/include/clang/Driver/Options.td | 6 +- clang/lib/Driver/Driver.cpp | 2 +- clang/lib/Driver/ToolChains/Clang.cpp | 2 +- clang/lib/Sema/SemaSYCL.cpp | 14 +++- .../anonymous_integration_footer.cpp | 75 ++++++++++--------- .../anonymous_integration_footer2.cpp | 21 +++--- clang/test/CodeGenSYCL/integration_footer.cpp | 22 +++--- clang/test/Driver/sycl-host-compiler.cpp | 4 +- clang/test/Driver/sycl-int-footer.cpp | 2 +- clang/test/Driver/sycl-intelfpga-aoco-win.cpp | 4 +- clang/test/Driver/sycl-intelfpga-aoco.cpp | 4 +- .../Driver/sycl-offload-intelfpga-emu.cpp | 4 +- clang/test/Driver/sycl-offload-intelfpga.cpp | 4 +- clang/test/Driver/sycl-offload-nvptx.cpp | 4 +- .../test/Driver/sycl-offload-static-lib-2.cpp | 2 +- clang/test/Driver/sycl-offload-static-lib.cpp | 2 +- clang/test/Driver/sycl-offload-win.c | 4 +- clang/test/Driver/sycl-offload-with-split.c | 34 ++++----- clang/test/Driver/sycl-offload.c | 70 ++++++++--------- clang/test/Driver/sycl-preprocess.cpp | 2 +- .../non_native/accelerator.cpp | 2 +- .../non_native/cpu.cpp | 2 +- .../non_native/cuda.cpp | 2 +- .../non_native/gpu.cpp | 2 +- 24 files changed, 149 insertions(+), 141 deletions(-) diff --git a/clang/include/clang/Driver/Options.td b/clang/include/clang/Driver/Options.td index bcd2b279841bf..2a20a367b9493 100644 --- a/clang/include/clang/Driver/Options.td +++ b/clang/include/clang/Driver/Options.td @@ -2610,9 +2610,9 @@ def fsycl_host_compiler_options_EQ : Joined<["-"], "fsycl-host-compiler-options= Flags<[CoreOption]>, HelpText<"When performing the host compilation with " "-fsycl-host-compiler specified, use the given options during that compile. " "Options are expected to be a quoted list of space separated options.">; -def fsycl_use_footer : Flag<["-"], "fsycl-use-footer">, Flags<[CoreOption]>, - HelpText<"Enable usage of the integration footer during SYCL enabled " - "compilations.">; +def fno_sycl_use_footer : Flag<["-"], "fno-sycl-use-footer">, Flags<[CoreOption]>, + HelpText<"Disable usage of the integration footer during SYCL enabled " + "compilations.">; def fsyntax_only : Flag<["-"], "fsyntax-only">, Flags<[NoXarchOption,CoreOption,CC1Option,FC1Option]>, Group; def ftabstop_EQ : Joined<["-"], "ftabstop=">, Group; diff --git a/clang/lib/Driver/Driver.cpp b/clang/lib/Driver/Driver.cpp index 68685d82c5503..ea68374b12fe1 100644 --- a/clang/lib/Driver/Driver.cpp +++ b/clang/lib/Driver/Driver.cpp @@ -5584,7 +5584,7 @@ Action *Driver::ConstructPhaseAction( if (Args.hasArg(options::OPT_verify_pch)) return C.MakeAction(Input, types::TY_Nothing); if (Args.hasArg(options::OPT_fsycl) && - Args.hasArg(options::OPT_fsycl_use_footer) && + !Args.hasArg(options::OPT_fno_sycl_use_footer) && TargetDeviceOffloadKind == Action::OFK_None) { // Performing a host compilation with -fsycl. Append the integrated // footer to the preprocessed source file. We then add another diff --git a/clang/lib/Driver/ToolChains/Clang.cpp b/clang/lib/Driver/ToolChains/Clang.cpp index 4dec8df6c5346..ff3b68292ebd2 100644 --- a/clang/lib/Driver/ToolChains/Clang.cpp +++ b/clang/lib/Driver/ToolChains/Clang.cpp @@ -4621,7 +4621,7 @@ void Clang::ConstructJob(Compilation &C, const JobAction &JA, CmdArgs.push_back(Args.MakeArgString(HeaderOpt)); } - if (Args.hasArg(options::OPT_fsycl_use_footer)) { + if (!Args.hasArg(options::OPT_fno_sycl_use_footer)) { // Add the integration footer option to generated the footer. StringRef Footer(D.getIntegrationFooter(Input.getBaseInput())); if (!Footer.empty()) { diff --git a/clang/lib/Sema/SemaSYCL.cpp b/clang/lib/Sema/SemaSYCL.cpp index 139b0a1db08c6..1e61f67b924b8 100644 --- a/clang/lib/Sema/SemaSYCL.cpp +++ b/clang/lib/Sema/SemaSYCL.cpp @@ -4752,6 +4752,10 @@ bool SYCLIntegrationFooter::emit(StringRef IntHeaderName) { } void SYCLIntegrationFooter::emitSpecIDName(raw_ostream &O, const VarDecl *VD) { + PrintingPolicy Policy{S.getLangOpts()}; + Policy.adjustForCPlusPlusFwdDecl(); + Policy.SuppressTypedefs = true; + Policy.SuppressUnwrittenScope = true; // FIXME: Figure out the spec-constant unique name here. // Note that this changes based on the linkage of the variable. // We typically want to use the __builtin_unique_stable_name for the variable @@ -4762,7 +4766,7 @@ void SYCLIntegrationFooter::emitSpecIDName(raw_ostream &O, const VarDecl *VD) { // ahead of it, so that we make sure it is unique across translation units. // This name should come from the yet implemented__builtin_unique_stable_name // feature that accepts variables and gives the mangling for that. - O << ""; + VD->printQualifiedName(O, Policy); } template @@ -4891,13 +4895,14 @@ bool SYCLIntegrationFooter::emit(raw_ostream &OS) { Policy.SuppressTypedefs = true; Policy.SuppressUnwrittenScope = true; + OS << "#include \n"; + // Used to uniquely name the 'shim's as we generate the names in each // anonymous namespace. unsigned ShimCounter = 0; for (const VarDecl *VD : SpecConstants) { VD = VD->getCanonicalDecl(); std::string TopShim = EmitSpecIdShims(OS, ShimCounter, VD); - OS << "#include \n"; OS << "__SYCL_INLINE_NAMESPACE(cl) {\n"; OS << "namespace sycl {\n"; OS << "namespace detail {\n"; @@ -4912,9 +4917,10 @@ bool SYCLIntegrationFooter::emit(raw_ostream &OS) { } OS << ">() {\n"; - OS << " return \""; + OS << " return " + "__builtin_unique_stable_name(specialization_id_name_generator<"; emitSpecIDName(OS, VD); - OS << "\";\n"; + OS << ">);\n"; OS << "}\n"; OS << "} // namespace detail\n"; OS << "} // namespace sycl\n"; diff --git a/clang/test/CodeGenSYCL/anonymous_integration_footer.cpp b/clang/test/CodeGenSYCL/anonymous_integration_footer.cpp index 7e0dd7d1e1972..1e49c032836ab 100644 --- a/clang/test/CodeGenSYCL/anonymous_integration_footer.cpp +++ b/clang/test/CodeGenSYCL/anonymous_integration_footer.cpp @@ -14,12 +14,13 @@ using namespace cl; // variable. struct S1 { static constexpr sycl::specialization_id a{1}; - // CHECK: __SYCL_INLINE_NAMESPACE(cl) { + // CHECK: #include + // CHECK-NEXT: __SYCL_INLINE_NAMESPACE(cl) { // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::S1::a>() { - // CHECK-NEXT: return ""; + // CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -31,7 +32,7 @@ constexpr sycl::specialization_id b{2}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::b>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -42,7 +43,7 @@ inline constexpr sycl::specialization_id c{3}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::c>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -53,7 +54,7 @@ static constexpr sycl::specialization_id d{4}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::d>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -62,23 +63,23 @@ static constexpr sycl::specialization_id d{4}; namespace { struct S2 { static constexpr sycl::specialization_id a{18}; - // CHECK-NEXT: namespace { - // CHECK-NEXT: namespace __sycl_detail { - // CHECK-NEXT: static constexpr decltype(S2::a) &__spec_id_shim_[[SHIM_ID:[0-9]+]]() { - // CHECK-NEXT: return S2::a; - // CHECK-NEXT: } - // CHECK-NEXT: } // namespace __sycl_detail - // CHECK-NEXT: } // namespace - // CHECK-NEXT: __SYCL_INLINE_NAMESPACE(cl) { - // CHECK-NEXT: namespace sycl { - // CHECK-NEXT: namespace detail { - // CHECK-NEXT: template<> - // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { -// CHECK-NEXT: return ""; - // CHECK-NEXT: } - // CHECK-NEXT: } // namespace detail - // CHECK-NEXT: } // namespace sycl - // CHECK-NEXT: } // __SYCL_INLINE_NAMESPACE(cl) +// CHECK-NEXT: namespace { +// CHECK-NEXT: namespace __sycl_detail { +// CHECK-NEXT: static constexpr decltype(S2::a) &__spec_id_shim_[[SHIM_ID:[0-9]+]]() { +// CHECK-NEXT: return S2::a; +// CHECK-NEXT: } +// CHECK-NEXT: } // namespace __sycl_detail +// CHECK-NEXT: } // namespace +// CHECK-NEXT: __SYCL_INLINE_NAMESPACE(cl) { +// CHECK-NEXT: namespace sycl { +// CHECK-NEXT: namespace detail { +// CHECK-NEXT: template<> +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); +// CHECK-NEXT: } +// CHECK-NEXT: } // namespace detail +// CHECK-NEXT: } // namespace sycl +// CHECK-NEXT: } // __SYCL_INLINE_NAMESPACE(cl) }; } // namespace @@ -92,7 +93,7 @@ template class S3<1>; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::S3<1>::a>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator::a>); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -103,7 +104,7 @@ template class S3<2>; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::S3<2>::a>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator::a>); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -116,7 +117,7 @@ constexpr sycl::specialization_id same_name{5}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::inner::same_name>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -128,7 +129,7 @@ constexpr sycl::specialization_id same_name{6}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::same_name>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -147,7 +148,7 @@ constexpr sycl::specialization_id same_name{7}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -168,7 +169,7 @@ constexpr sycl::specialization_id same_name{8}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -192,7 +193,7 @@ constexpr sycl::specialization_id same_name{9}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::inner::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -207,7 +208,7 @@ constexpr sycl::specialization_id same_name{10}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::outer::same_name>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -228,7 +229,7 @@ constexpr sycl::specialization_id same_name{11}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::outer::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -250,7 +251,7 @@ constexpr sycl::specialization_id same_name{12}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::outer::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -288,7 +289,7 @@ constexpr sycl::specialization_id same_name{13}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::outer::__sycl_detail::__spec_id_shim_[[SHIM_ID_2]]()>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -313,7 +314,7 @@ constexpr sycl::specialization_id same_name{14}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -343,7 +344,7 @@ constexpr sycl::specialization_id same_name{15}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::__sycl_detail::__spec_id_shim_[[SHIM_ID2]]()>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -373,7 +374,7 @@ constexpr sycl::specialization_id same_name{16}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::__sycl_detail::__spec_id_shim_[[SHIM_ID2]]()>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -391,7 +392,7 @@ constexpr sycl::specialization_id same_name{17}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::outer::inner::same_name>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl diff --git a/clang/test/CodeGenSYCL/anonymous_integration_footer2.cpp b/clang/test/CodeGenSYCL/anonymous_integration_footer2.cpp index 77c4e01ee5897..51c4b03d29f30 100644 --- a/clang/test/CodeGenSYCL/anonymous_integration_footer2.cpp +++ b/clang/test/CodeGenSYCL/anonymous_integration_footer2.cpp @@ -13,12 +13,13 @@ using namespace cl; struct S1 { static constexpr sycl::specialization_id a{1}; }; -// CHECK: __SYCL_INLINE_NAMESPACE(cl) { +// CHECK: #include +// CHECK-NEXT: __SYCL_INLINE_NAMESPACE(cl) { // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::S1::a>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -30,7 +31,7 @@ constexpr sycl::specialization_id b{202}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::b>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -41,7 +42,7 @@ inline constexpr sycl::specialization_id c{3}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::c>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -52,7 +53,7 @@ static constexpr sycl::specialization_id d{205}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::d>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -65,7 +66,7 @@ constexpr sycl::specialization_id same_name{5}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::inner::same_name>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -77,7 +78,7 @@ constexpr sycl::specialization_id same_name{6}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::same_name>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -96,7 +97,7 @@ constexpr sycl::specialization_id same_name{207}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -117,7 +118,7 @@ constexpr sycl::specialization_id same_name{208}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -144,7 +145,7 @@ constexpr sycl::specialization_id same_name{209}; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::outer::inner::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl diff --git a/clang/test/CodeGenSYCL/integration_footer.cpp b/clang/test/CodeGenSYCL/integration_footer.cpp index a4694940dc8ff..5622a4dcdff83 100644 --- a/clang/test/CodeGenSYCL/integration_footer.cpp +++ b/clang/test/CodeGenSYCL/integration_footer.cpp @@ -14,7 +14,7 @@ cl::sycl::specialization_id GlobalSpecID; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::GlobalSpecID>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -25,7 +25,7 @@ struct Wrapper { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::Wrapper::WrapperSpecID>() { - // CHECK-NEXT: return ""; + // CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -40,7 +40,7 @@ template class WrapperTemplate; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::WrapperTemplate::WrapperSpecID>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator::WrapperSpecID>); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -49,7 +49,7 @@ template class WrapperTemplate; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::WrapperTemplate::WrapperSpecID>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator::WrapperSpecID>); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -60,7 +60,7 @@ specialization_id NSSpecID; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::Foo::NSSpecID>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -70,7 +70,7 @@ specialization_id InlineNSSpecID; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::Foo::InlineNSSpecID>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -79,7 +79,7 @@ specialization_id NSSpecID; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::Foo::Bar::NSSpecID>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -90,7 +90,7 @@ struct Wrapper { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::Foo::Wrapper::WrapperSpecID>() { - // CHECK-NEXT: return ""; + // CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -105,7 +105,7 @@ template class WrapperTemplate; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::Foo::WrapperTemplate::WrapperSpecID>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator::WrapperSpecID>); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -114,7 +114,7 @@ template class WrapperTemplate; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::Foo::WrapperTemplate::WrapperSpecID>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator::WrapperSpecID>); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl @@ -136,7 +136,7 @@ specialization_id AnonNSSpecID; // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::Foo::__sycl_detail::__spec_id_shim_[[SHIM0]]()>() { -// CHECK-NEXT: return ""; +// CHECK-NEXT: return __builtin_unique_stable_name(specialization_id_name_generator); // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl diff --git a/clang/test/Driver/sycl-host-compiler.cpp b/clang/test/Driver/sycl-host-compiler.cpp index ba2bb3c73f717..563b1de168362 100644 --- a/clang/test/Driver/sycl-host-compiler.cpp +++ b/clang/test/Driver/sycl-host-compiler.cpp @@ -2,14 +2,14 @@ // REQUIRES: clang-driver /// enabling with -fsycl-host-compiler -// RUN: %clangxx -fsycl-use-footer -fsycl -fsycl-host-compiler=/some/dir/g++ %s -### 2>&1 \ +// RUN: %clangxx -fsycl -fsycl-host-compiler=/some/dir/g++ %s -### 2>&1 \ // RUN: | FileCheck -check-prefix=HOST_COMPILER %s // HOST_COMPILER: clang{{.*}} "-fsycl-is-device"{{.*}} "-fsycl-int-header=[[INTHEADER:.+\.h]]" "-fsycl-int-footer={{.*}}" // HOST_COMPILER: g++{{.*}} "-E" "-I" "{{.*}}bin{{[/\\]+}}..{{[/\\]+}}include{{[/\\]+}}sycl"{{.*}} "-o" "[[TMPII:.+\.ii]]"{{.*}} "-include" "[[INTHEADER]]" // HOST_COMPILER: g++{{.*}} "-I" "{{.*}}bin{{[/\\]+}}..{{[/\\]+}}include{{[/\\]+}}sycl"{{.*}} "-o" "[[HOSTOBJ:.+\.o]]"{{.*}} // HOST_COMPILER: ld{{.*}} "[[HOSTOBJ]]" -// RUN: %clang_cl -fsycl-use-footer -fsycl -fsycl-host-compiler=/some/dir/cl %s -### 2>&1 \ +// RUN: %clang_cl -fsycl -fsycl-host-compiler=/some/dir/cl %s -### 2>&1 \ // RUN: | FileCheck -check-prefix=HOST_COMPILER_CL %s // HOST_COMPILER_CL: clang{{.*}} "-fsycl-is-device"{{.*}} "-fsycl-int-header=[[INTHEADER:.+\.h]]" "-fsycl-int-footer={{.*}}" // HOST_COMPILER_CL: cl{{.*}} "-P" "-Fi[[TMPII:.+\.ii]]"{{.*}} "-I" "{{.*}}bin{{[/\\]+}}..{{[/\\]+}}include{{[/\\]+}}sycl"{{.*}} "-FI" "[[INTHEADER]]" diff --git a/clang/test/Driver/sycl-int-footer.cpp b/clang/test/Driver/sycl-int-footer.cpp index 2036d16512479..d57c91ab7c55e 100644 --- a/clang/test/Driver/sycl-int-footer.cpp +++ b/clang/test/Driver/sycl-int-footer.cpp @@ -1,5 +1,5 @@ /// Check compilation tool steps when using the integrated footer -// RUN: %clangxx -fsycl -fsycl-use-footer %s -### 2>&1 \ +// RUN: %clangxx -fsycl %s -### 2>&1 \ // RUN: | FileCheck -check-prefix FOOTER %s // FOOTER: clang{{.*}} "-fsycl-is-device"{{.*}} "-fsycl-int-header=[[INTHEADER:.+\.h]]" "-fsycl-int-footer=[[INTFOOTER:.+\h]]" "-sycl-std={{.*}}" // FOOTER: clang{{.*}} "-include" "[[INTHEADER]]"{{.*}} "-fsycl-is-host"{{.*}} "-E"{{.*}} "-o" "[[PREPROC:.+\.ii]]" diff --git a/clang/test/Driver/sycl-intelfpga-aoco-win.cpp b/clang/test/Driver/sycl-intelfpga-aoco-win.cpp index 300740183ab98..e6352958abcd4 100755 --- a/clang/test/Driver/sycl-intelfpga-aoco-win.cpp +++ b/clang/test/Driver/sycl-intelfpga-aoco-win.cpp @@ -7,9 +7,9 @@ // RUN: clang-offload-wrapper -o %t-aoco.bc -host=x86_64-pc-windows-msvc -kind=sycl -target=fpga_aoco-intel-unknown-sycldevice %t.aoco // RUN: llc -filetype=obj -o %t-aoco.o %t-aoco.bc // RUN: llvm-ar crv %t_aoco.a %t.o %t2.o %t-aoco.o -// RUN: %clang_cl -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fintelfpga -foffload-static-lib=%t_aoco.a %s -ccc-print-phases 2>&1 \ +// RUN: %clang_cl -fsycl -fno-sycl-device-lib=all -fintelfpga -foffload-static-lib=%t_aoco.a %s -ccc-print-phases 2>&1 \ // RUN: | FileCheck -check-prefixes=CHK-FPGA-AOCO-PHASES-WIN %s -// RUN: %clangxx -fsycl-use-footer -target x86_64-pc-windows-msvc -fsycl -fno-sycl-device-lib=all -fintelfpga -foffload-static-lib=%t_aoco.a %s -ccc-print-phases 2>&1 \ +// RUN: %clangxx -target x86_64-pc-windows-msvc -fsycl -fno-sycl-device-lib=all -fintelfpga -foffload-static-lib=%t_aoco.a %s -ccc-print-phases 2>&1 \ // RUN: | FileCheck -check-prefixes=CHK-FPGA-AOCO-PHASES-WIN %s // CHK-FPGA-AOCO-PHASES-WIN: 0: input, "[[INPUTA:.+\.a]]", object, (host-sycl) // CHK-FPGA-AOCO-PHASES-WIN: 1: input, "[[INPUTSRC:.+\.cpp]]", c++, (host-sycl) diff --git a/clang/test/Driver/sycl-intelfpga-aoco.cpp b/clang/test/Driver/sycl-intelfpga-aoco.cpp index 18075dded5e03..9df70a0f926b9 100755 --- a/clang/test/Driver/sycl-intelfpga-aoco.cpp +++ b/clang/test/Driver/sycl-intelfpga-aoco.cpp @@ -13,7 +13,7 @@ // RUN: llc -filetype=obj -o %t-aoco_cl.o %t-aoco_cl.bc // RUN: llvm-ar crv %t_aoco.a %t.o %t2.o %t-aoco.o // RUN: llvm-ar crv %t_aoco_cl.a %t.o %t2_cl.o %t-aoco_cl.o -// RUN: %clangxx -fsycl-use-footer -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fintelfpga %t_aoco.a %s -ccc-print-phases 2>&1 \ +// RUN: %clangxx -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fintelfpga %t_aoco.a %s -ccc-print-phases 2>&1 \ // RUN: | FileCheck -check-prefix=CHK-FPGA-AOCO-PHASES %s // CHK-FPGA-AOCO-PHASES: 0: input, "[[INPUTA:.+\.a]]", object, (host-sycl) // CHK-FPGA-AOCO-PHASES: 1: input, "[[INPUTCPP:.+\.cpp]]", c++, (host-sycl) @@ -46,7 +46,7 @@ // CHK-FPGA-AOCO-PHASES: 28: offload, "host-sycl (x86_64-unknown-linux-gnu)" {12}, "device-sycl (spir64_fpga-unknown-unknown-sycldevice)" {27}, image /// FPGA AOCO Windows phases check -// RUN: %clang_cl -fsycl -fsycl-use-footer -fno-sycl-device-lib=all -fintelfpga -foffload-static-lib=%t_aoco_cl.a %s -ccc-print-phases 2>&1 \ +// RUN: %clang_cl -fsycl -fno-sycl-device-lib=all -fintelfpga -foffload-static-lib=%t_aoco_cl.a %s -ccc-print-phases 2>&1 \ // RUN: | FileCheck -check-prefixes=CHK-FPGA-AOCO-PHASES-WIN %s // CHK-FPGA-AOCO-PHASES-WIN: 0: input, "[[INPUTA:.+\.a]]", object, (host-sycl) // CHK-FPGA-AOCO-PHASES-WIN: 1: input, "[[INPUTSRC:.+\.cpp]]", c++, (host-sycl) diff --git a/clang/test/Driver/sycl-offload-intelfpga-emu.cpp b/clang/test/Driver/sycl-offload-intelfpga-emu.cpp index 8c06abd321a17..c64062d3d55f0 100644 --- a/clang/test/Driver/sycl-offload-intelfpga-emu.cpp +++ b/clang/test/Driver/sycl-offload-intelfpga-emu.cpp @@ -180,9 +180,9 @@ /// -fintelfpga -fsycl-link from source // RUN: touch %t.cpp -// RUN: %clangxx -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fintelfpga -fsycl-link=early %t.cpp -ccc-print-phases 2>&1 \ +// RUN: %clangxx -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fintelfpga -fsycl-link=early %t.cpp -ccc-print-phases 2>&1 \ // RUN: | FileCheck -check-prefixes=CHK-FPGA-LINK-SRC %s -// RUN: %clang_cl --target=x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fintelfpga -fsycl-link=early %t.cpp -ccc-print-phases 2>&1 \ +// RUN: %clang_cl --target=x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fintelfpga -fsycl-link=early %t.cpp -ccc-print-phases 2>&1 \ // RUN: | FileCheck -check-prefixes=CHK-FPGA-LINK-SRC %s // CHK-FPGA-LINK-SRC: 0: input, "[[INPUT:.+\.cpp]]", c++, (host-sycl) // CHK-FPGA-LINK-SRC: 1: preprocessor, {0}, c++-cpp-output, (host-sycl) diff --git a/clang/test/Driver/sycl-offload-intelfpga.cpp b/clang/test/Driver/sycl-offload-intelfpga.cpp index afabdc2bcd793..5efb7a931489e 100644 --- a/clang/test/Driver/sycl-offload-intelfpga.cpp +++ b/clang/test/Driver/sycl-offload-intelfpga.cpp @@ -242,9 +242,9 @@ /// -fintelfpga -fsycl-link from source // RUN: touch %t.cpp -// RUN: %clangxx -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fintelfpga -fsycl-link=early %t.cpp -ccc-print-phases 2>&1 \ +// RUN: %clangxx -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fintelfpga -fsycl-link=early %t.cpp -ccc-print-phases 2>&1 \ // RUN: | FileCheck -check-prefixes=CHK-FPGA-LINK-SRC %s -// RUN: %clang_cl --target=x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fintelfpga -fsycl-link=early %t.cpp -ccc-print-phases 2>&1 \ +// RUN: %clang_cl --target=x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fintelfpga -fsycl-link=early %t.cpp -ccc-print-phases 2>&1 \ // RUN: | FileCheck -check-prefixes=CHK-FPGA-LINK-SRC %s // CHK-FPGA-LINK-SRC: 0: input, "[[INPUT:.+\.cpp]]", c++, (host-sycl) // CHK-FPGA-LINK-SRC: 1: preprocessor, {0}, c++-cpp-output, (host-sycl) diff --git a/clang/test/Driver/sycl-offload-nvptx.cpp b/clang/test/Driver/sycl-offload-nvptx.cpp index 4e23a1859fc86..ed5a3051781f8 100644 --- a/clang/test/Driver/sycl-offload-nvptx.cpp +++ b/clang/test/Driver/sycl-offload-nvptx.cpp @@ -13,7 +13,7 @@ // CHK-ACTIONS: clang-offload-wrapper"{{.*}} "-host=x86_64-unknown-linux-gnu" "-target=nvptx64" "-kind=sycl"{{.*}} /// Check phases w/out specifying a compute capability. -// RUN: %clangxx -ccc-print-phases -std=c++11 -target x86_64-unknown-linux-gnu -fsycl -fsycl-use-footer \ +// RUN: %clangxx -ccc-print-phases -std=c++11 -target x86_64-unknown-linux-gnu -fsycl \ // RUN: -fsycl-targets=nvptx64-nvidia-nvcl-sycldevice %s 2>&1 \ // RUN: | FileCheck -check-prefix=CHK-PHASES-NO-CC %s // CHK-PHASES-NO-CC: 0: input, "{{.*}}", c++, (host-sycl) @@ -35,7 +35,7 @@ // CHK-PHASES-NO-CC: 16: offload, "host-sycl (x86_64-unknown-linux-gnu)" {11}, "device-sycl (nvptx64-nvidia-nvcl-sycldevice:sm_50)" {15}, image /// Check phases specifying a compute capability. -// RUN: %clangxx -ccc-print-phases -std=c++11 -target x86_64-unknown-linux-gnu -fsycl -fsycl-use-footer \ +// RUN: %clangxx -ccc-print-phases -std=c++11 -target x86_64-unknown-linux-gnu -fsycl \ // RUN: -fsycl-targets=nvptx64-nvidia-nvcl-sycldevice \ // RUN: -Xsycl-target-backend "--cuda-gpu-arch=sm_35" %s 2>&1 \ // RUN: | FileCheck -check-prefix=CHK-PHASES %s diff --git a/clang/test/Driver/sycl-offload-static-lib-2.cpp b/clang/test/Driver/sycl-offload-static-lib-2.cpp index 1ca26c0d5673f..940fac6ff6221 100644 --- a/clang/test/Driver/sycl-offload-static-lib-2.cpp +++ b/clang/test/Driver/sycl-offload-static-lib-2.cpp @@ -45,7 +45,7 @@ /// test behaviors of fat static lib from source // RUN: touch %t_lib.a -// RUN: %clangxx -target x86_64-unknown-linux-gnu -fsycl-use-footer -fno-sycl-device-lib=all -fsycl %t_lib.a -ccc-print-phases %s 2>&1 \ +// RUN: %clangxx -target x86_64-unknown-linux-gnu -fno-sycl-device-lib=all -fsycl %t_lib.a -ccc-print-phases %s 2>&1 \ // RUN: | FileCheck %s -check-prefix=STATIC_LIB_SRC // STATIC_LIB_SRC: 0: input, "[[INPUTA:.+\.a]]", object, (host-sycl) // STATIC_LIB_SRC: 1: input, "[[INPUTC:.+\.cpp]]", c++, (host-sycl) diff --git a/clang/test/Driver/sycl-offload-static-lib.cpp b/clang/test/Driver/sycl-offload-static-lib.cpp index 6b90554c231ae..5d63da239dbdd 100644 --- a/clang/test/Driver/sycl-offload-static-lib.cpp +++ b/clang/test/Driver/sycl-offload-static-lib.cpp @@ -48,7 +48,7 @@ /// test behaviors of -foffload-static-lib= from source // RUN: touch %t.a -// RUN: %clangxx -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -foffload-static-lib=%t.a -ccc-print-phases %s 2>&1 \ +// RUN: %clangxx -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -foffload-static-lib=%t.a -ccc-print-phases %s 2>&1 \ // RUN: | FileCheck %s -check-prefix=FOFFLOAD_STATIC_LIB_SRC // FOFFLOAD_STATIC_LIB_SRC: 0: input, "[[INPUTA:.+\.a]]", object, (host-sycl) diff --git a/clang/test/Driver/sycl-offload-win.c b/clang/test/Driver/sycl-offload-win.c index b61aff44c4428..acddd39b87365 100644 --- a/clang/test/Driver/sycl-offload-win.c +++ b/clang/test/Driver/sycl-offload-win.c @@ -57,9 +57,9 @@ /// Test behaviors of -foffload-static-lib= from source. // RUN: touch %t-orig.lib -// RUN: %clang --target=x86_64-pc-windows-msvc -fsycl-use-footer -fsycl -fno-sycl-device-lib=all %t-orig.lib -ccc-print-phases %s 2>&1 \ +// RUN: %clang --target=x86_64-pc-windows-msvc -fsycl -fno-sycl-device-lib=all %t-orig.lib -ccc-print-phases %s 2>&1 \ // RUN: | FileCheck %s -check-prefix=FOFFLOAD_STATIC_LIB_SRC -// RUN: %clang_cl --target=x86_64-pc-windows-msvc -fsycl-use-footer -fsycl -fno-sycl-device-lib=all %t-orig.lib -ccc-print-phases %s 2>&1 \ +// RUN: %clang_cl --target=x86_64-pc-windows-msvc -fsycl -fno-sycl-device-lib=all %t-orig.lib -ccc-print-phases %s 2>&1 \ // RUN: | FileCheck %s -check-prefix=FOFFLOAD_STATIC_LIB_SRC // FOFFLOAD_STATIC_LIB_SRC: 0: input, "[[INPUTLIB:.+\.lib]]", object, (host-sycl) // FOFFLOAD_STATIC_LIB_SRC: 1: input, "[[INPUTC:.+\.c]]", c++, (host-sycl) diff --git a/clang/test/Driver/sycl-offload-with-split.c b/clang/test/Driver/sycl-offload-with-split.c index 612639e505741..6d7b621685c7d 100644 --- a/clang/test/Driver/sycl-offload-with-split.c +++ b/clang/test/Driver/sycl-offload-with-split.c @@ -12,17 +12,17 @@ /// preprocessor and another one joining the device linking outputs to the host /// action. The same graph should be generated when no -fsycl-targets is used /// The same phase graph will be used with -fsycl-use-bitcode -// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split -fsycl-targets=spir64-unknown-unknown-sycldevice %s 2>&1 \ +// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split -fsycl-targets=spir64-unknown-unknown-sycldevice %s 2>&1 \ // RUN: | FileCheck -check-prefixes=CHK-PHASES,CHK-PHASES-DEFAULT-MODE %s -// RUN: %clang_cl -ccc-print-phases --target=x86_64-pc-windows-msvc -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split=per_source -fsycl-targets=spir64-unknown-unknown-sycldevice %s 2>&1 \ +// RUN: %clang_cl -ccc-print-phases --target=x86_64-pc-windows-msvc -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split=per_source -fsycl-targets=spir64-unknown-unknown-sycldevice %s 2>&1 \ // RUN: | FileCheck -check-prefixes=CHK-PHASES,CHK-PHASES-CL-MODE %s -// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split=per_source -fno-sycl-use-bitcode %s 2>&1 \ +// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split=per_source -fno-sycl-use-bitcode %s 2>&1 \ // RUN: | FileCheck -check-prefixes=CHK-PHASES,CHK-PHASES-DEFAULT-MODE %s -// RUN: %clang_cl -ccc-print-phases --target=x86_64-pc-windows-msvc -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split=per_source -fno-sycl-use-bitcode %s 2>&1 \ +// RUN: %clang_cl -ccc-print-phases --target=x86_64-pc-windows-msvc -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split=per_source -fno-sycl-use-bitcode %s 2>&1 \ // RUN: | FileCheck -check-prefixes=CHK-PHASES,CHK-PHASES-CL-MODE %s -// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split=per_source -fsycl-use-bitcode %s 2>&1 \ +// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split=per_source -fsycl-use-bitcode %s 2>&1 \ // RUN: | FileCheck -check-prefixes=CHK-PHASES,CHK-PHASES-DEFAULT-MODE %s -// RUN: %clang_cl -ccc-print-phases --target=x86_64-pc-windows-msvc -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split=per_source -fsycl-use-bitcode %s 2>&1 \ +// RUN: %clang_cl -ccc-print-phases --target=x86_64-pc-windows-msvc -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split=per_source -fsycl-use-bitcode %s 2>&1 \ // RUN: | FileCheck -check-prefixes=CHK-PHASES,CHK-PHASES-CL-MODE %s // CHK-PHASES: 0: input, "[[INPUT:.+\.c]]", c++, (host-sycl) // CHK-PHASES: 1: preprocessor, {0}, c++-cpp-output, (host-sycl) @@ -50,7 +50,7 @@ /// Check the phases also add a library to make sure it is treated as input by /// the device. -// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -lsomelib -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split -fsycl-targets=spir64-unknown-unknown-sycldevice %s 2>&1 \ +// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -lsomelib -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split -fsycl-targets=spir64-unknown-unknown-sycldevice %s 2>&1 \ // RUN: | FileCheck -check-prefix=CHK-PHASES-LIB %s // CHK-PHASES-LIB: 0: input, "somelib", object, (host-sycl) // CHK-PHASES-LIB: 1: input, "[[INPUT:.+\.c]]", c++, (host-sycl) @@ -77,7 +77,7 @@ /// Check the phases when using and multiple source files // RUN: echo " " > %t.c -// RUN: %clang -ccc-print-phases -lsomelib -target x86_64-unknown-linux-gnu -fsycl -fsycl-use-footer -fno-sycl-device-lib=all -fsycl-device-code-split -fsycl-targets=spir64-unknown-unknown-sycldevice %s %t.c 2>&1 \ +// RUN: %clang -ccc-print-phases -lsomelib -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split -fsycl-targets=spir64-unknown-unknown-sycldevice %s %t.c 2>&1 \ // RUN: | FileCheck -check-prefix=CHK-PHASES-FILES %s // CHK-PHASES-FILES: 0: input, "somelib", object, (host-sycl) @@ -138,7 +138,7 @@ /// Check separate compilation with offloading - unbundling with source // RUN: touch %t.o -// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -lsomelib -fsycl -fsycl-use-footer -fno-sycl-device-lib=all -fsycl-device-code-split %t.o -fsycl-targets=spir64-unknown-unknown-sycldevice %s 2>&1 \ +// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -lsomelib -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split %t.o -fsycl-targets=spir64-unknown-unknown-sycldevice %s 2>&1 \ // RUN: | FileCheck -check-prefix=CHK-UBUACTIONS %s // CHK-UBUACTIONS: 0: input, "somelib", object, (host-sycl) // CHK-UBUACTIONS: 1: input, "[[INPUT1:.+\.o]]", object, (host-sycl) @@ -166,11 +166,11 @@ /// ########################################################################### /// Ahead of Time compilation for fpga, gen, cpu -// RUN: %clang -target x86_64-unknown-linux-gnu -ccc-print-phases -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split -fsycl-targets=spir64_fpga-unknown-unknown-sycldevice %s 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -ccc-print-phases -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split -fsycl-targets=spir64_fpga-unknown-unknown-sycldevice %s 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-PHASES-AOT,CHK-PHASES-FPGA -// RUN: %clang -target x86_64-unknown-linux-gnu -ccc-print-phases -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split -fsycl-targets=spir64_gen-unknown-unknown-sycldevice %s 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -ccc-print-phases -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split -fsycl-targets=spir64_gen-unknown-unknown-sycldevice %s 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-PHASES-AOT,CHK-PHASES-GEN -// RUN: %clang -target x86_64-unknown-linux-gnu -ccc-print-phases -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split -fsycl-targets=spir64_x86_64-unknown-unknown-sycldevice %s 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -ccc-print-phases -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split -fsycl-targets=spir64_x86_64-unknown-unknown-sycldevice %s 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-PHASES-AOT,CHK-PHASES-CPU // CHK-PHASES-AOT: 0: input, "[[INPUT:.+\.c]]", c++, (host-sycl) // CHK-PHASES-AOT: 1: preprocessor, {0}, c++-cpp-output, (host-sycl) @@ -200,13 +200,13 @@ /// ########################################################################### /// Ahead of Time compilation for fpga, gen, cpu - tool invocation -// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split -fsycl-targets=spir64_fpga-unknown-unknown-sycldevice -Xshardware %s -### 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split -fsycl-targets=spir64_fpga-unknown-unknown-sycldevice -Xshardware %s -### 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-TOOLS-AOT,CHK-TOOLS-FPGA -// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split -fintelfpga -Xshardware %s -### 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split -fintelfpga -Xshardware %s -### 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-TOOLS-AOT,CHK-TOOLS-FPGA -// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split -fsycl-targets=spir64_gen-unknown-unknown-sycldevice %s -### 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split -fsycl-targets=spir64_gen-unknown-unknown-sycldevice %s -### 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-TOOLS-AOT,CHK-TOOLS-GEN -// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split -fsycl-targets=spir64_x86_64-unknown-unknown-sycldevice %s -### 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split -fsycl-targets=spir64_x86_64-unknown-unknown-sycldevice %s -### 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-TOOLS-AOT,CHK-TOOLS-CPU // CHK-TOOLS-AOT: clang{{.*}} "-fsycl-is-device"{{.*}} "-fsycl-int-header=[[INPUT1:.+\-header.+\.h]]" "-fsycl-int-footer={{.*}}"{{.*}} "-o" "[[OUTPUT1:.+\.bc]]" // CHK-TOOLS-AOT: llvm-link{{.*}} "[[OUTPUT1]]" "-o" "[[OUTPUT2:.+\.bc]]" @@ -228,7 +228,7 @@ /// ########################################################################### /// offload with multiple targets, including AOT -// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split -fsycl-targets=spir64-unknown-unknown-sycldevice,spir64_fpga-unknown-unknown-sycldevice,spir64_gen-unknown-unknown-sycldevice -ccc-print-phases %s 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fsycl-device-code-split -fsycl-targets=spir64-unknown-unknown-sycldevice,spir64_fpga-unknown-unknown-sycldevice,spir64_gen-unknown-unknown-sycldevice -ccc-print-phases %s 2>&1 \ // RUN: | FileCheck -check-prefix=CHK-PHASE-MULTI-TARG %s // CHK-PHASE-MULTI-TARG: 0: input, "[[INPUT:.+\.c]]", c++, (host-sycl) // CHK-PHASE-MULTI-TARG: 1: preprocessor, {0}, c++-cpp-output, (host-sycl) diff --git a/clang/test/Driver/sycl-offload.c b/clang/test/Driver/sycl-offload.c index d028dbcfcf428..01877910477f5 100644 --- a/clang/test/Driver/sycl-offload.c +++ b/clang/test/Driver/sycl-offload.c @@ -161,17 +161,17 @@ /// preprocessor and another one joining the device linking outputs to the host /// action. The same graph should be generated when no -fsycl-targets is used /// The same phase graph will be used with -fsycl-use-bitcode -// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fsycl-targets=spir64-unknown-unknown-sycldevice -fno-sycl-device-lib=all %s 2>&1 \ +// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -fsycl -fsycl-targets=spir64-unknown-unknown-sycldevice -fno-sycl-device-lib=all %s 2>&1 \ // RUN: | FileCheck -check-prefixes=CHK-PHASES,CHK-PHASES-DEFAULT-MODE %s -// RUN: %clang_cl -ccc-print-phases -fsycl-use-footer -fsycl -fsycl-targets=spir64-unknown-unknown-sycldevice -fno-sycl-device-lib=all %s 2>&1 \ +// RUN: %clang_cl -ccc-print-phases -fsycl -fsycl-targets=spir64-unknown-unknown-sycldevice -fno-sycl-device-lib=all %s 2>&1 \ // RUN: | FileCheck -check-prefixes=CHK-PHASES,CHK-PHASES-CL-MODE %s -// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-use-bitcode -fno-sycl-device-lib=all %s 2>&1 \ +// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-use-bitcode -fno-sycl-device-lib=all %s 2>&1 \ // RUN: | FileCheck -check-prefixes=CHK-PHASES,CHK-PHASES-DEFAULT-MODE %s -// RUN: %clang_cl -ccc-print-phases -fsycl-use-footer -fsycl -fno-sycl-use-bitcode -fno-sycl-device-lib=all %s 2>&1 \ +// RUN: %clang_cl -ccc-print-phases -fsycl -fno-sycl-use-bitcode -fno-sycl-device-lib=all %s 2>&1 \ // RUN: | FileCheck -check-prefixes=CHK-PHASES,CHK-PHASES-CL-MODE %s -// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fsycl-use-bitcode -fno-sycl-device-lib=all %s 2>&1 \ +// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -fsycl -fsycl-use-bitcode -fno-sycl-device-lib=all %s 2>&1 \ // RUN: | FileCheck -check-prefixes=CHK-PHASES,CHK-PHASES-DEFAULT-MODE %s -// RUN: %clang_cl -ccc-print-phases -fsycl-use-footer -fsycl -fsycl-use-bitcode -fno-sycl-device-lib=all %s 2>&1 \ +// RUN: %clang_cl -ccc-print-phases -fsycl -fsycl-use-bitcode -fno-sycl-device-lib=all %s 2>&1 \ // RUN: | FileCheck -check-prefixes=CHK-PHASES,CHK-PHASES-CL-MODE %s // CHK-PHASES: 0: input, "[[INPUT:.+\.c]]", c++, (host-sycl) // CHK-PHASES: 1: preprocessor, {0}, c++-cpp-output, (host-sycl) @@ -198,7 +198,7 @@ /// ########################################################################### /// Check the compilation flow to verify that the integrated header is filtered -// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -c %s -### 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl -c %s -### 2>&1 \ // RUN: | FileCheck %s -check-prefix=CHK-INT-HEADER // CHK-INT-HEADER: clang{{.*}} "-fsycl-is-device"{{.*}} "-fsycl-int-header=[[INPUT1:.+\-header.+\.h]]" "-fsycl-int-footer={{.*}}"{{.*}} "-o" "[[OUTPUT1:.+\.bc]]" // CHK-INT-HEADER: clang{{.*}} "-triple" "x86_64-unknown-linux-gnu" {{.*}} "-include" "[[INPUT1]]" "-dependency-filter" "[[INPUT1]]" {{.*}} "-o" "[[TMPII:.+\.ii]]" @@ -209,7 +209,7 @@ /// Check the phases also add a library to make sure it is treated as input by /// the device. -// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -lsomelib -fsycl-use-footer -fsycl -fsycl-targets=spir64-unknown-unknown-sycldevice -fno-sycl-device-lib=all %s 2>&1 \ +// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -lsomelib -fsycl -fsycl-targets=spir64-unknown-unknown-sycldevice -fno-sycl-device-lib=all %s 2>&1 \ // RUN: | FileCheck -check-prefix=CHK-PHASES-LIB %s // CHK-PHASES-LIB: 0: input, "somelib", object, (host-sycl) // CHK-PHASES-LIB: 1: input, "[[INPUT:.+\.c]]", c++, (host-sycl) @@ -243,7 +243,7 @@ /// Check the phases when using and multiple source files // RUN: echo " " > %t.c -// RUN: %clang -ccc-print-phases -lsomelib -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fsycl-targets=spir64-unknown-unknown-sycldevice -fno-sycl-device-lib=all %s %t.c 2>&1 \ +// RUN: %clang -ccc-print-phases -lsomelib -target x86_64-unknown-linux-gnu -fsycl -fsycl-targets=spir64-unknown-unknown-sycldevice -fno-sycl-device-lib=all %s %t.c 2>&1 \ // RUN: | FileCheck -check-prefix=CHK-PHASES-FILES %s // CHK-PHASES-FILES: 0: input, "somelib", object, (host-sycl) @@ -281,7 +281,7 @@ /// ########################################################################### /// Check separate compilation with offloading - bundling actions -// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -c -o %t.o -lsomelib -fsycl-targets=spir64-unknown-unknown-sycldevice %s 2>&1 \ +// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -fsycl -c -o %t.o -lsomelib -fsycl-targets=spir64-unknown-unknown-sycldevice %s 2>&1 \ // RUN: | FileCheck -check-prefix=CHK-BUACTIONS %s // CHK-BUACTIONS: 0: input, "[[INPUT:.+\.c]]", c++, (device-sycl) // CHK-BUACTIONS: 1: preprocessor, {0}, c++-cpp-output, (device-sycl) @@ -323,7 +323,7 @@ /// Check separate compilation with offloading - unbundling with source // RUN: touch %t.o -// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -lsomelib -fsycl-use-footer -fsycl -fno-sycl-device-lib=all %t.o -fsycl-targets=spir64-unknown-unknown-sycldevice %s 2>&1 \ +// RUN: %clang -ccc-print-phases -target x86_64-unknown-linux-gnu -lsomelib -fsycl -fno-sycl-device-lib=all %t.o -fsycl-targets=spir64-unknown-unknown-sycldevice %s 2>&1 \ // RUN: | FileCheck -check-prefix=CHK-UBUACTIONS %s // CHK-UBUACTIONS: 0: input, "somelib", object, (host-sycl) // CHK-UBUACTIONS: 1: input, "[[INPUT1:.+\.o]]", object, (host-sycl) @@ -467,7 +467,7 @@ /// Check offload with multiple triples, multiple binaries passed through -fsycl-add-targets -// RUN: %clang -### -ccc-print-phases -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -o %t.out -fsycl-add-targets=spir64-unknown-unknown-sycldevice:dummy.spv,spir64_fpga-unknown-unknown-sycldevice:dummy.aocx,spir64_gen-unknown-unknown-sycldevice:dummy_Gen9core.bin %s 2>&1 \ +// RUN: %clang -### -ccc-print-phases -target x86_64-unknown-linux-gnu -fsycl -o %t.out -fsycl-add-targets=spir64-unknown-unknown-sycldevice:dummy.spv,spir64_fpga-unknown-unknown-sycldevice:dummy.aocx,spir64_gen-unknown-unknown-sycldevice:dummy_Gen9core.bin %s 2>&1 \ // RUN: | FileCheck -check-prefix=CHK-ADD-TARGETS-MUL %s // CHK-ADD-TARGETS-MUL: 0: input, "[[INPUT:.+\.c]]", c++, (host-sycl) // CHK-ADD-TARGETS-MUL: 1: preprocessor, {0}, c++-cpp-output, (host-sycl) @@ -493,7 +493,7 @@ /// Check offload with single triple, multiple binaries passed through -fsycl-add-targets -// RUN: %clang -### -ccc-print-phases -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -o %t.out -fsycl-add-targets=spir64-unknown-unknown-sycldevice:dummy0.spv,spir64-unknown-unknown-sycldevice:dummy1.spv,spir64-unknown-unknown-sycldevice:dummy2.spv %s 2>&1 \ +// RUN: %clang -### -ccc-print-phases -target x86_64-unknown-linux-gnu -fsycl -o %t.out -fsycl-add-targets=spir64-unknown-unknown-sycldevice:dummy0.spv,spir64-unknown-unknown-sycldevice:dummy1.spv,spir64-unknown-unknown-sycldevice:dummy2.spv %s 2>&1 \ // RUN: | FileCheck -check-prefix=CHK-ADD-TARGETS-MUL-BINS %s // CHK-ADD-TARGETS-MUL-BINS: 0: input, "[[INPUT:.+\.c]]", c++, (host-sycl) // CHK-ADD-TARGETS-MUL-BINS: 1: preprocessor, {0}, c++-cpp-output, (host-sycl) @@ -519,7 +519,7 @@ /// Check regular offload with an additional AOT binary passed through -fsycl-add-targets (same triple) -// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64-unknown-unknown-sycldevice -fsycl-add-targets=spir64-unknown-unknown-sycldevice:dummy.spv -ccc-print-phases %s 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64-unknown-unknown-sycldevice -fsycl-add-targets=spir64-unknown-unknown-sycldevice:dummy.spv -ccc-print-phases %s 2>&1 \ // RUN: | FileCheck -check-prefix=CHK-ADD-TARGETS-REG %s // CHK-ADD-TARGETS-REG: 0: input, "[[INPUT:.+\.c]]", c++, (host-sycl) // CHK-ADD-TARGETS-REG: 1: preprocessor, {0}, c++-cpp-output, (host-sycl) @@ -546,7 +546,7 @@ /// ########################################################################### /// Check regular offload with multiple additional AOT binaries passed through -fsycl-add-targets -// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64-unknown-unknown-sycldevice -fsycl-add-targets=spir64_fpga-unknown-unknown-sycldevice:dummy.aocx,spir64_gen-unknown-unknown-sycldevice:dummy_Gen9core.bin,spir64_x86_64-unknown-unknown-sycldevice:dummy.ir -ccc-print-phases %s 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64-unknown-unknown-sycldevice -fsycl-add-targets=spir64_fpga-unknown-unknown-sycldevice:dummy.aocx,spir64_gen-unknown-unknown-sycldevice:dummy_Gen9core.bin,spir64_x86_64-unknown-unknown-sycldevice:dummy.ir -ccc-print-phases %s 2>&1 \ // RUN: | FileCheck -check-prefix=CHK-ADD-TARGETS-REG-MUL %s // CHK-ADD-TARGETS-REG-MUL: 0: input, "[[INPUT:.+\.c]]", c++, (host-sycl) // CHK-ADD-TARGETS-REG-MUL: 1: preprocessor, {0}, c++-cpp-output, (host-sycl) @@ -624,11 +624,11 @@ /// ########################################################################### /// Ahead of Time compilation for fpga, gen, cpu -// RUN: %clang -target x86_64-unknown-linux-gnu -ccc-print-phases -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_fpga-unknown-unknown-sycldevice %s 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -ccc-print-phases -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_fpga-unknown-unknown-sycldevice %s 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-PHASES-AOT,CHK-PHASES-FPGA -// RUN: %clang -target x86_64-unknown-linux-gnu -ccc-print-phases -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_gen-unknown-unknown-sycldevice %s 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -ccc-print-phases -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_gen-unknown-unknown-sycldevice %s 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-PHASES-AOT,CHK-PHASES-GEN -// RUN: %clang -target x86_64-unknown-linux-gnu -ccc-print-phases -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_x86_64-unknown-unknown-sycldevice %s 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -ccc-print-phases -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_x86_64-unknown-unknown-sycldevice %s 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-PHASES-AOT,CHK-PHASES-CPU // CHK-PHASES-AOT: 0: input, "[[INPUT:.+\.c]]", c++, (host-sycl) // CHK-PHASES-AOT: 1: preprocessor, {0}, c++-cpp-output, (host-sycl) @@ -660,29 +660,29 @@ /// ########################################################################### /// Ahead of Time compilation for fpga, gen, cpu - tool invocation -// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_fpga-unknown-unknown-sycldevice %s -### 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_fpga-unknown-unknown-sycldevice %s -### 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-TOOLS-AOT,CHK-TOOLS-FPGA,CHK-TOOLS-FPGA-EMU -// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fintelfpga %s -### 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fintelfpga %s -### 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-TOOLS-AOT,CHK-TOOLS-FPGA,CHK-TOOLS-FPGA-EMU -// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_fpga-unknown-unknown-sycldevice -Xshardware %s -### 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_fpga-unknown-unknown-sycldevice -Xshardware %s -### 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-TOOLS-AOT,CHK-TOOLS-FPGA,CHK-TOOLS-FPGA-HW -// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fintelfpga -Xshardware %s -### 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fintelfpga -Xshardware %s -### 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-TOOLS-AOT,CHK-TOOLS-FPGA,CHK-TOOLS-FPGA-HW -// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_fpga-unknown-unknown-sycldevice -Xssimulation %s -### 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_fpga-unknown-unknown-sycldevice -Xssimulation %s -### 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-TOOLS-AOT,CHK-TOOLS-FPGA,CHK-TOOLS-FPGA-HW -// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fintelfpga -Xssimulation %s -### 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fintelfpga -Xssimulation %s -### 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-TOOLS-AOT,CHK-TOOLS-FPGA,CHK-TOOLS-FPGA-HW -// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_fpga-unknown-unknown-sycldevice -Xsemulator %s -### 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_fpga-unknown-unknown-sycldevice -Xsemulator %s -### 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-TOOLS-AOT,CHK-TOOLS-FPGA,CHK-TOOLS-FPGA-EMU -// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fintelfpga -Xsemulator %s -### 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fintelfpga -Xsemulator %s -### 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-TOOLS-AOT,CHK-TOOLS-FPGA,CHK-TOOLS-FPGA-EMU -// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_gen-unknown-unknown-sycldevice %s -### 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_gen-unknown-unknown-sycldevice %s -### 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-TOOLS-AOT,CHK-TOOLS-GEN -// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_x86_64-unknown-unknown-sycldevice %s -### 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_x86_64-unknown-unknown-sycldevice %s -### 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-TOOLS-AOT,CHK-TOOLS-CPU -// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_gen-unknown-unknown-sycldevice %s -### 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_gen-unknown-unknown-sycldevice %s -### 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-TOOLS-AOT,CHK-TOOLS-GEN -// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_x86_64-unknown-unknown-sycldevice %s -### 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64_x86_64-unknown-unknown-sycldevice %s -### 2>&1 \ // RUN: | FileCheck %s -check-prefixes=CHK-TOOLS-AOT,CHK-TOOLS-CPU // CHK-TOOLS-FPGA: clang{{.*}} "-triple" "spir64_fpga-unknown-unknown-sycldevice" // CHK-TOOLS-GEN: clang{{.*}} "-triple" "spir64_gen-unknown-unknown-sycldevice" @@ -813,7 +813,7 @@ /// ########################################################################### /// offload with multiple targets, including AOT -// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64-unknown-unknown-sycldevice,spir64_fpga-unknown-unknown-sycldevice,spir64_gen-unknown-unknown-sycldevice -ccc-print-phases %s 2>&1 \ +// RUN: %clang -target x86_64-unknown-linux-gnu -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64-unknown-unknown-sycldevice,spir64_fpga-unknown-unknown-sycldevice,spir64_gen-unknown-unknown-sycldevice -ccc-print-phases %s 2>&1 \ // RUN: | FileCheck -check-prefix=CHK-PHASE-MULTI-TARG %s // CHK-PHASE-MULTI-TARG: 0: input, "[[INPUT:.+\.c]]", c++, (host-sycl) // CHK-PHASE-MULTI-TARG: 1: preprocessor, {0}, c++-cpp-output, (host-sycl) @@ -857,9 +857,9 @@ /// ########################################################################### /// Verify that -save-temps does not crash -// RUN: %clang -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -target x86_64-unknown-linux-gnu -save-temps %s -### 2>&1 -// RUN: %clang -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64-unknown-unknown-sycldevice -target x86_64-unknown-linux-gnu -save-temps %s -### 2>&1 -// RUN: %clangxx -fsycl-use-footer -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64-unknown-unknown-sycldevice -target x86_64-unknown-linux-gnu -save-temps %s -### 2>&1 \ +// RUN: %clang -fsycl -fno-sycl-device-lib=all -target x86_64-unknown-linux-gnu -save-temps %s -### 2>&1 +// RUN: %clang -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64-unknown-unknown-sycldevice -target x86_64-unknown-linux-gnu -save-temps %s -### 2>&1 +// RUN: %clangxx -fsycl -fno-sycl-device-lib=all -fsycl-targets=spir64-unknown-unknown-sycldevice -target x86_64-unknown-linux-gnu -save-temps %s -### 2>&1 \ // RUN: | FileCheck %s --check-prefixes=CHK-FSYCL-SAVE-TEMPS,CHK-FSYCL-SAVE-TEMPS-CONFL // CHK-FSYCL-SAVE-TEMPS: clang{{.*}} "-fsycl-is-device"{{.*}} "-o" "[[DEVICE_BASE_NAME:[a-z0-9-]+]].ii" // CHK-FSYCL-SAVE-TEMPS: clang{{.*}} "-fsycl-is-device"{{.*}} "-fsycl-int-header=[[HEADER_NAME:.+\-header.+\.h]]" "-fsycl-int-footer={{.*}}"{{.*}} "-o" "[[DEVICE_BASE_NAME]].bc"{{.*}} "[[DEVICE_BASE_NAME]].ii" @@ -880,7 +880,7 @@ // CHK-FSYCL-SAVE-TEMPS: ld{{.*}} "[[HOST_BASE_NAME]].o"{{.*}} "[[DEVICE_OBJ_NAME]]" /// -fsycl with /Fo testing -// RUN: %clang_cl -fsycl-use-footer -fsycl /Fosomefile.obj -c %s -### 2>&1 \ +// RUN: %clang_cl -fsycl /Fosomefile.obj -c %s -### 2>&1 \ // RUN: | FileCheck -check-prefix=FO-CHECK %s // FO-CHECK: clang{{.*}} "-fsycl-int-header=[[HEADER:.+\.h]]" "-fsycl-int-footer={{.*}}"{{.*}} "-o" "[[OUTPUT1:.+\.bc]]" // FO-CHECK: clang{{.*}} "-include" "[[HEADER]]" {{.*}} "-o" "[[TMPII:.+\.ii]]" diff --git a/clang/test/Driver/sycl-preprocess.cpp b/clang/test/Driver/sycl-preprocess.cpp index 11a7bd7d836dc..0f9159cb0414e 100644 --- a/clang/test/Driver/sycl-preprocess.cpp +++ b/clang/test/Driver/sycl-preprocess.cpp @@ -6,7 +6,7 @@ // RUN: %clang_cl -fsycl -P -Fi%t_output.ii %s -### 2>&1 \ // RUN: | FileCheck -check-prefix PREPROC_ONLY %s // PREPROC_ONLY: clang{{.*}} "-fsycl-is-device"{{.*}} "-E"{{.*}} "-o" "[[DEVICE_OUT:.+\.ii]]" -// PREPROC_ONLY: clang{{.*}} "-fsycl-is-device"{{.*}} "-fsycl-int-header=[[INTHEADER:.+\.h]]"{{.*}} "-fsyntax-only" +// PREPROC_ONLY: clang{{.*}} "-fsycl-is-device"{{.*}} "-fsycl-int-header=[[INTHEADER:.+\.h]]" "-fsycl-int-footer=[[INTFOOTER:.+\.h]]"{{.*}} "-fsyntax-only" // PREPROC_ONLY: clang{{.*}} "-include" "[[INTHEADER]]"{{.*}} "-fsycl-is-host"{{.*}} "-o" "[[HOST_OUT:.+\.ii]]" // PREPROC_ONLY: clang-offload-bundler{{.*}} "-type=ii"{{.*}} "-outputs={{.+_output.ii}}" "-inputs=[[DEVICE_OUT]],[[HOST_OUT]]" diff --git a/sycl/test/on-device/basic_tests/specialization_constants/non_native/accelerator.cpp b/sycl/test/on-device/basic_tests/specialization_constants/non_native/accelerator.cpp index 081b7c3242cea..c11958d10d07a 100644 --- a/sycl/test/on-device/basic_tests/specialization_constants/non_native/accelerator.cpp +++ b/sycl/test/on-device/basic_tests/specialization_constants/non_native/accelerator.cpp @@ -1,6 +1,6 @@ // REQUIRES: aoc, accelerator -// RUN: %clangxx -fsycl -fsycl-use-footer -fsycl-targets=spir64_fpga-unknown-unknown-sycldevice %S/Inputs/common.cpp -o %t.out +// RUN: %clangxx -fsycl -fsycl-targets=spir64_fpga-unknown-unknown-sycldevice %S/Inputs/common.cpp -o %t.out // RUN: %ACC_RUN_PLACEHOLDER %t.out // This test checks correctness of SYCL2020 non-native specialization constants diff --git a/sycl/test/on-device/basic_tests/specialization_constants/non_native/cpu.cpp b/sycl/test/on-device/basic_tests/specialization_constants/non_native/cpu.cpp index 555f7801bcd75..a2c91f1b58de3 100644 --- a/sycl/test/on-device/basic_tests/specialization_constants/non_native/cpu.cpp +++ b/sycl/test/on-device/basic_tests/specialization_constants/non_native/cpu.cpp @@ -1,6 +1,6 @@ // REQUIRES: opencl-aot, cpu -// RUN: %clangxx -fsycl -fsycl-use-footer -fsycl-targets=spir64_x86_64-unknown-unknown-sycldevice %S/Inputs/common.cpp -o %t.out +// RUN: %clangxx -fsycl -fsycl-targets=spir64_x86_64-unknown-unknown-sycldevice %S/Inputs/common.cpp -o %t.out // RUN: %CPU_RUN_PLACEHOLDER %t.out // This test checks correctness of SYCL2020 non-native specialization constants diff --git a/sycl/test/on-device/basic_tests/specialization_constants/non_native/cuda.cpp b/sycl/test/on-device/basic_tests/specialization_constants/non_native/cuda.cpp index 5229c3c2d5fc4..01d0d37e04f26 100644 --- a/sycl/test/on-device/basic_tests/specialization_constants/non_native/cuda.cpp +++ b/sycl/test/on-device/basic_tests/specialization_constants/non_native/cuda.cpp @@ -1,6 +1,6 @@ // REQUIRES: cuda -// RUN: %clangxx -fsycl -fsycl-use-footer -fsycl-targets=nvptx64-unknown-unknown-sycldevice %S/Inputs/common.cpp -o %t.out +// RUN: %clangxx -fsycl -fsycl-targets=nvptx64-unknown-unknown-sycldevice %S/Inputs/common.cpp -o %t.out // RUN: env SYCL_DEVICE_FILTER=cuda %t.out // TODO: enable this test then compile-time error in sycl-post-link is fixed diff --git a/sycl/test/on-device/basic_tests/specialization_constants/non_native/gpu.cpp b/sycl/test/on-device/basic_tests/specialization_constants/non_native/gpu.cpp index 84e56e51bc7a5..df9f0a7937dc3 100644 --- a/sycl/test/on-device/basic_tests/specialization_constants/non_native/gpu.cpp +++ b/sycl/test/on-device/basic_tests/specialization_constants/non_native/gpu.cpp @@ -2,7 +2,7 @@ // UNSUPPORTED: cuda // CUDA is not compatible with SPIR. -// RUN: %clangxx -fsycl -fsycl-use-footer -fsycl-targets=spir64_gen-unknown-unknown-sycldevice -Xsycl-target-backend=spir64_gen-unknown-unknown-sycldevice "-device *" %S/Inputs/common.cpp -o %t.out +// RUN: %clangxx -fsycl -fsycl-targets=spir64_gen-unknown-unknown-sycldevice -Xsycl-target-backend=spir64_gen-unknown-unknown-sycldevice "-device *" %S/Inputs/common.cpp -o %t.out // RUN: %GPU_RUN_PLACEHOLDER %t.out // This test checks correctness of SYCL2020 non-native specialization constants From 0d3e0009df40a4513392ead0b7b45f2a2e41034a Mon Sep 17 00:00:00 2001 From: Dmitry Vodopyanov Date: Wed, 9 Jun 2021 16:46:59 +0300 Subject: [PATCH 06/21] Fix some test failures --- clang/lib/Sema/SemaSYCL.cpp | 5 ++--- sycl/include/CL/sycl/detail/spec_const_integration.hpp | 2 +- 2 files changed, 3 insertions(+), 4 deletions(-) diff --git a/clang/lib/Sema/SemaSYCL.cpp b/clang/lib/Sema/SemaSYCL.cpp index ff8fe2b3a6006..b382b9a571683 100644 --- a/clang/lib/Sema/SemaSYCL.cpp +++ b/clang/lib/Sema/SemaSYCL.cpp @@ -5029,6 +5029,7 @@ bool SYCLIntegrationFooter::emit(raw_ostream &OS) { Policy.SuppressUnwrittenScope = true; OS << "#include \n"; + OS << "#include \n"; // Used to uniquely name the 'shim's as we generate the names in each // anonymous namespace. @@ -5051,7 +5052,7 @@ bool SYCLIntegrationFooter::emit(raw_ostream &OS) { OS << ">() {\n"; OS << " return " - "__builtin_unique_stable_name(specialization_id_name_generator<"; + "__builtin_sycl_unique_stable_name(specialization_id_name_generator<"; emitSpecIDName(OS, VD); OS << ">);\n"; OS << "}\n"; @@ -5059,8 +5060,6 @@ bool SYCLIntegrationFooter::emit(raw_ostream &OS) { OS << "} // namespace sycl\n"; OS << "} // __SYCL_INLINE_NAMESPACE(cl)\n"; } - - OS << "#include \n"; return true; } diff --git a/sycl/include/CL/sycl/detail/spec_const_integration.hpp b/sycl/include/CL/sycl/detail/spec_const_integration.hpp index e25fe8c549bae..69e2c76197913 100644 --- a/sycl/include/CL/sycl/detail/spec_const_integration.hpp +++ b/sycl/include/CL/sycl/detail/spec_const_integration.hpp @@ -21,7 +21,7 @@ template struct specialization_id_name_generator {}; // Translates SYCL 2020 specialization constant type to its name. template const char *get_spec_constant_symbolic_ID() { #ifdef SYCL_LANGUAGE_VERSION - return __builtin_unique_stable_name( + return __builtin_sycl_unique_stable_name( specialization_id_name_generator); #else return ""; From ed1686dcbebc05c17a8642dee5cb582e0660cf7f Mon Sep 17 00:00:00 2001 From: Dmitry Vodopyanov Date: Tue, 15 Jun 2021 12:55:22 +0300 Subject: [PATCH 07/21] Update PR to use new builtin --- clang/lib/Driver/ToolChains/Clang.cpp | 2 +- sycl/include/CL/sycl/detail/kernel_desc.hpp | 8 -------- sycl/include/CL/sycl/detail/spec_const_integration.hpp | 5 +---- sycl/include/CL/sycl/kernel_handler.hpp | 6 ++---- 4 files changed, 4 insertions(+), 17 deletions(-) diff --git a/clang/lib/Driver/ToolChains/Clang.cpp b/clang/lib/Driver/ToolChains/Clang.cpp index c449a097beaad..140af3abf630e 100644 --- a/clang/lib/Driver/ToolChains/Clang.cpp +++ b/clang/lib/Driver/ToolChains/Clang.cpp @@ -1175,7 +1175,7 @@ void Clang::AddPreprocessingOptions(Compilation &C, const JobAction &JA, // When preprocessing using the integration footer, add the comments // to the first preprocessing step. if (JA.isOffloading(Action::OFK_SYCL) && !ContainsAppendFooterAction(&JA) && - Args.hasArg(options::OPT_fsycl_use_footer) && + !Args.hasArg(options::OPT_fno_sycl_use_footer) && JA.isDeviceOffloading(Action::OFK_None)) CmdArgs.push_back("-C"); diff --git a/sycl/include/CL/sycl/detail/kernel_desc.hpp b/sycl/include/CL/sycl/detail/kernel_desc.hpp index 6ed7b415b5212..bf766e1d1deb7 100644 --- a/sycl/include/CL/sycl/detail/kernel_desc.hpp +++ b/sycl/include/CL/sycl/detail/kernel_desc.hpp @@ -17,14 +17,6 @@ __SYCL_INLINE_NAMESPACE(cl) { namespace sycl { namespace detail { -// This guard is needed because the libsycl.so can be compiled with C++ <=14 -// while the code requires C++17. This code is not supposed to be used by the -// libsycl.so so it should not be a problem. -#if __cplusplus > 201402L -// Definition is in spec_const_integration.hpp -template struct specialization_id_name_generator; -#endif - #ifndef __SYCL_DEVICE_ONLY__ #define _Bool bool #endif diff --git a/sycl/include/CL/sycl/detail/spec_const_integration.hpp b/sycl/include/CL/sycl/detail/spec_const_integration.hpp index 69e2c76197913..6025d35a5950d 100644 --- a/sycl/include/CL/sycl/detail/spec_const_integration.hpp +++ b/sycl/include/CL/sycl/detail/spec_const_integration.hpp @@ -16,13 +16,10 @@ namespace sycl { namespace detail { #if __cplusplus >= 201703L -template struct specialization_id_name_generator {}; - // Translates SYCL 2020 specialization constant type to its name. template const char *get_spec_constant_symbolic_ID() { #ifdef SYCL_LANGUAGE_VERSION - return __builtin_sycl_unique_stable_name( - specialization_id_name_generator); + return __builtin_sycl_unique_stable_id(SpecName); #else return ""; #endif diff --git a/sycl/include/CL/sycl/kernel_handler.hpp b/sycl/include/CL/sycl/kernel_handler.hpp index 9a1fbaf15efb6..26d164944330f 100644 --- a/sycl/include/CL/sycl/kernel_handler.hpp +++ b/sycl/include/CL/sycl/kernel_handler.hpp @@ -49,8 +49,7 @@ class kernel_handler { typename T = typename std::remove_reference_t::value_type, std::enable_if_t> * = nullptr> T getSpecializationConstantOnDevice() { - const char *SymbolicID = __builtin_sycl_unique_stable_name( - detail::specialization_id_name_generator); + const char *SymbolicID = __builtin_sycl_unique_stable_id(S); return __sycl_getScalar2020SpecConstantValue( SymbolicID, &S, MSpecializationConstantsBuffer); } @@ -59,8 +58,7 @@ class kernel_handler { typename T = typename std::remove_reference_t::value_type, std::enable_if_t> * = nullptr> T getSpecializationConstantOnDevice() { - const char *SymbolicID = __builtin_sycl_unique_stable_name( - detail::specialization_id_name_generator); + const char *SymbolicID = __builtin_sycl_unique_stable_id(S); return __sycl_getComposite2020SpecConstantValue( SymbolicID, &S, MSpecializationConstantsBuffer); } From feb54c4b1e459090ec8f23169966a76f872aceaa Mon Sep 17 00:00:00 2001 From: Dmitry Vodopyanov Date: Tue, 15 Jun 2021 13:58:08 +0300 Subject: [PATCH 08/21] Fix SYCL :: basic_tests/SYCL-2020*** tests --- .../SYCL-2020-spec-const-ids-order.cpp | 16 ++++++------- .../basic_tests/SYCL-2020-spec-constants.cpp | 24 +++++++++---------- 2 files changed, 20 insertions(+), 20 deletions(-) diff --git a/sycl/test/basic_tests/SYCL-2020-spec-const-ids-order.cpp b/sycl/test/basic_tests/SYCL-2020-spec-const-ids-order.cpp index b71d70defd625..ecb8b1c6850ba 100644 --- a/sycl/test/basic_tests/SYCL-2020-spec-const-ids-order.cpp +++ b/sycl/test/basic_tests/SYCL-2020-spec-const-ids-order.cpp @@ -46,13 +46,13 @@ int main() { } // CHECK-PROP: [SYCL/specialization constants] -// CHECK-PROP-NEXT: _ZTSN2cl4sycl6detail32specialization_id_name_generatorIL_ZL5Val23EEE -// CHECK-PROP-NEXT: _ZTSN2cl4sycl6detail32specialization_id_name_generatorIL_ZL10ConstantIdEEE -// CHECK-PROP-NEXT: _ZTSN2cl4sycl6detail32specialization_id_name_generatorIL_ZL11SecondValueEEE -// CHECK-PROP-NEXT: _ZTSN2cl4sycl6detail32specialization_id_name_generatorIL_ZL11SpecConst42EEE +// CHECK-PROP-NEXT: [[UNIQUE_PREFIX:[a-z0-9]+]]____ZL5Val23 +// CHECK-PROP-NEXT: [[UNIQUE_PREFIX]]____ZL10ConstantId +// CHECK-PROP-NEXT: [[UNIQUE_PREFIX]]____ZL11SecondValue +// CHECK-PROP-NEXT: [[UNIQUE_PREFIX]]____ZL11SpecConst42 // // CHECK-IR: !sycl.specialization-constants = !{![[#MD0:]], ![[#MD1:]], ![[#MD2:]], ![[#MD3:]]} -// CHECK-IR: ![[#MD0]] = !{!"_ZTSN2cl4sycl6detail32specialization_id_name_generatorIL_ZL5Val23EEE", i32 [[#ID:]] -// CHECK-IR: ![[#MD1]] = !{!"_ZTSN2cl4sycl6detail32specialization_id_name_generatorIL_ZL10ConstantIdEEE", i32 [[#ID+1]] -// CHECK-IR: ![[#MD2]] = !{!"_ZTSN2cl4sycl6detail32specialization_id_name_generatorIL_ZL11SecondValueEEE", i32 [[#ID+2]] -// CHECK-IR: ![[#MD3]] = !{!"_ZTSN2cl4sycl6detail32specialization_id_name_generatorIL_ZL11SpecConst42EEE", i32 [[#ID+3]] +// CHECK-IR: ![[#MD0]] = !{!"[[UNIQUE_PREFIX:[a-z0-9]+]]____ZL5Val23", i32 [[#ID:]] +// CHECK-IR: ![[#MD1]] = !{!"[[UNIQUE_PREFIX]]____ZL10ConstantId", i32 [[#ID+1]] +// CHECK-IR: ![[#MD2]] = !{!"[[UNIQUE_PREFIX]]____ZL11SecondValue", i32 [[#ID+2]] +// CHECK-IR: ![[#MD3]] = !{!"[[UNIQUE_PREFIX]]____ZL11SpecConst42", i32 [[#ID+3]] diff --git a/sycl/test/basic_tests/SYCL-2020-spec-constants.cpp b/sycl/test/basic_tests/SYCL-2020-spec-constants.cpp index 1db6b8f2587da..83b9e854e98dd 100644 --- a/sycl/test/basic_tests/SYCL-2020-spec-constants.cpp +++ b/sycl/test/basic_tests/SYCL-2020-spec-constants.cpp @@ -86,19 +86,19 @@ int main() { } // CHECK: [SYCL/specialization constants] -// CHECK-DAG: _ZTSN2cl4sycl6detail32specialization_id_name_generatorIL_ZL12composite_idEEE=2| +// CHECK-DAG: [[UNIQUE_PREFIX:[a-z0-9]+]]____ZL12composite_id=2| // See FIXME above about bool type support -// CHECK-disabled: _ZTSN2cl4sycl6detail32specialization_id_name_generatorIL_ZL7bool_idEEE=2| -// CHECK-DAG: _ZTSN2cl4sycl6detail32specialization_id_name_generatorIL_ZL7int8_idEEE=2| -// CHECK-DAG: _ZTSN2cl4sycl6detail32specialization_id_name_generatorIL_ZL8float_idEEE=2| -// CHECK-DAG: _ZTSN2cl4sycl6detail32specialization_id_name_generatorIL_ZL8int16_idEEE=2| -// CHECK-DAG: _ZTSN2cl4sycl6detail32specialization_id_name_generatorIL_ZL8int32_idEEE=2| -// CHECK-DAG: _ZTSN2cl4sycl6detail32specialization_id_name_generatorIL_ZL8int64_idEEE=2| -// CHECK-DAG: _ZTSN2cl4sycl6detail32specialization_id_name_generatorIL_ZL8uint8_idEEE=2| -// CHECK-DAG: _ZTSN2cl4sycl6detail32specialization_id_name_generatorIL_ZL9double_idEEE=2| -// CHECK-DAG: _ZTSN2cl4sycl6detail32specialization_id_name_generatorIL_ZL9uint16_idEEE=2| -// CHECK-DAG: _ZTSN2cl4sycl6detail32specialization_id_name_generatorIL_ZL9uint32_idEEE=2| -// CHECK-DAG: _ZTSN2cl4sycl6detail32specialization_id_name_generatorIL_ZL9uint64_idEEE=2| +// CHECK-disabled: [[UNIQUE_PREFIX]]____IL_ZL7bool_id=2| +// CHECK-DAG: [[UNIQUE_PREFIX]]____ZL7int8_id=2| +// CHECK-DAG: [[UNIQUE_PREFIX]]____ZL8float_id=2| +// CHECK-DAG: [[UNIQUE_PREFIX]]____ZL8int16_id=2| +// CHECK-DAG: [[UNIQUE_PREFIX]]____ZL8int32_id=2| +// CHECK-DAG: [[UNIQUE_PREFIX]]____ZL8int64_id=2| +// CHECK-DAG: [[UNIQUE_PREFIX]]____ZL8uint8_id=2| +// CHECK-DAG: [[UNIQUE_PREFIX]]____ZL9double_id=2| +// CHECK-DAG: [[UNIQUE_PREFIX]]____ZL9uint16_id=2| +// CHECK-DAG: [[UNIQUE_PREFIX]]____ZL9uint32_id=2| +// CHECK-DAG: [[UNIQUE_PREFIX]]____ZL9uint64_id=2| // FIXME: check line for half constant // CHECK-RT-NOT: [SYCL/specialization constants default values] From e56ee417374054e0bcccc8574192cf51601722cf Mon Sep 17 00:00:00 2001 From: Alexey Sachkov Date: Wed, 16 Jun 2021 16:00:26 +0300 Subject: [PATCH 09/21] Fix check-clang --- clang/test/CodeGenSYCL/anonymous_integration_footer.cpp | 9 +++++---- clang/test/CodeGenSYCL/anonymous_integration_footer2.cpp | 9 +++++---- clang/test/CodeGenSYCL/integration_footer.cpp | 5 +++-- clang/test/Driver/sycl-host-compiler.cpp | 8 ++++++-- clang/test/Driver/sycl-int-footer.cpp | 4 ++-- clang/test/Driver/sycl-offload-win.c | 2 ++ clang/test/Driver/sycl-preprocess.cpp | 8 ++++++-- 7 files changed, 29 insertions(+), 16 deletions(-) diff --git a/clang/test/CodeGenSYCL/anonymous_integration_footer.cpp b/clang/test/CodeGenSYCL/anonymous_integration_footer.cpp index 479a4dfa5786b..8af6745e9bef1 100644 --- a/clang/test/CodeGenSYCL/anonymous_integration_footer.cpp +++ b/clang/test/CodeGenSYCL/anonymous_integration_footer.cpp @@ -9,13 +9,16 @@ int main() { cl::sycl::kernel_single_task([]() {}); } +// CHECK: #include +// CHECK-NEXT: #include + using namespace cl; + // Example ways in which the application can declare a "specialization_id" // variable. struct S1 { static constexpr sycl::specialization_id a{1}; - // CHECK: #include - // CHECK-NEXT: __SYCL_INLINE_NAMESPACE(cl) { + // CHECK: __SYCL_INLINE_NAMESPACE(cl) { // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> @@ -399,5 +402,3 @@ constexpr sycl::specialization_id same_name{17}; // CHECK-NEXT: } // __SYCL_INLINE_NAMESPACE(cl) } } // namespace outer - -// CHECK: #include diff --git a/clang/test/CodeGenSYCL/anonymous_integration_footer2.cpp b/clang/test/CodeGenSYCL/anonymous_integration_footer2.cpp index 9cd7aac03e67d..304f2a259b95c 100644 --- a/clang/test/CodeGenSYCL/anonymous_integration_footer2.cpp +++ b/clang/test/CodeGenSYCL/anonymous_integration_footer2.cpp @@ -8,13 +8,16 @@ int main() { cl::sycl::kernel_single_task([]() {}); } + +// CHECK: #include +// CHECK-NEXT: #include + using namespace cl; struct S1 { static constexpr sycl::specialization_id a{1}; }; -// CHECK: #include -// CHECK-NEXT: __SYCL_INLINE_NAMESPACE(cl) { +// CHECK: __SYCL_INLINE_NAMESPACE(cl) { // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> @@ -152,5 +155,3 @@ constexpr sycl::specialization_id same_name{209}; // CHECK-NEXT: } // __SYCL_INLINE_NAMESPACE(cl) } } - -// CHECK: #include diff --git a/clang/test/CodeGenSYCL/integration_footer.cpp b/clang/test/CodeGenSYCL/integration_footer.cpp index f3dfea126b8ed..5a06918ef42c4 100644 --- a/clang/test/CodeGenSYCL/integration_footer.cpp +++ b/clang/test/CodeGenSYCL/integration_footer.cpp @@ -7,6 +7,9 @@ int main() { cl::sycl::kernel_single_task([]() {}); } +// CHECK: #include +// CHECK-NEXT: #include + using namespace cl::sycl; cl::sycl::specialization_id GlobalSpecID; @@ -158,5 +161,3 @@ struct container { }; // CHECK-NOT: ::GetThing // CHECK-NOT: ::container::Thing - -// CHECK: #include diff --git a/clang/test/Driver/sycl-host-compiler.cpp b/clang/test/Driver/sycl-host-compiler.cpp index 563b1de168362..3e5d83742f7f8 100644 --- a/clang/test/Driver/sycl-host-compiler.cpp +++ b/clang/test/Driver/sycl-host-compiler.cpp @@ -29,12 +29,16 @@ // RUN: %clangxx -fsycl -fsycl-host-compiler=g++ -E %s -### 2>&1 \ // RUN: | FileCheck -check-prefix=HOST_PREPROCESS %s // HOST_PREPROCESS: g++{{.*}} "-E"{{.*}} "-o" "[[PPOUT:.+\.ii]]" -// HOST_PREPROCESS: clang-offload-bundler{{.*}} "-inputs={{.*}}.ii,[[PPOUT]]" +// HOST_PREPROCESS: append-file{{.*}} "[[PPOUT]]"{{.*}} "--output=[[APPEND:.+\.cpp]]" +// HOST_PREPROCESS: g++{{.*}} "[[APPEND]]"{{.*}} "-E"{{.*}} "-o" "[[PPOUT2:.+\.ii]]" +// HOST_PREPROCESS: clang-offload-bundler{{.*}} "-inputs={{.*}}.ii,[[PPOUT2]]" // RUN: %clang_cl -fsycl -fsycl-host-compiler=cl -E %s -### 2>&1 \ // RUN: | FileCheck -check-prefix=HOST_PREPROCESS_CL %s // HOST_PREPROCESS_CL: cl{{.*}} "-P"{{.*}} "-Fi[[PPOUT:.+\.ii]]" -// HOST_PREPROCESS_CL: clang-offload-bundler{{.*}} "-inputs={{.*}}.ii,[[PPOUT]]" +// HOST_PREPROCESS_CL: append-file{{.*}} "[[PPOUT]]"{{.*}} "--output=[[APPEND:.+\.cpp]]" +// HOST_PREPROCESS_CL: cl{{.*}} "[[APPEND]]"{{.*}} "-P"{{.*}} "-Fi[[PPOUT2:.+\.ii]]" +// HOST_PREPROCESS_CL: clang-offload-bundler{{.*}} "-inputs={{.*}}.ii,[[PPOUT2]]" /// obj output // RUN: %clangxx -fsycl -fsycl-host-compiler=g++ -c %s -### 2>&1 \ diff --git a/clang/test/Driver/sycl-int-footer.cpp b/clang/test/Driver/sycl-int-footer.cpp index 7dc3cafa8866e..7a7fc10499c84 100644 --- a/clang/test/Driver/sycl-int-footer.cpp +++ b/clang/test/Driver/sycl-int-footer.cpp @@ -8,7 +8,7 @@ // FOOTER-NOT: "-include" "[[INTHEADER]]" /// Preprocessed file creation with integration footer -// RUN: %clangxx -fsycl -fsycl-use-footer -E %s -### 2>&1 \ +// RUN: %clangxx -fsycl -E %s -### 2>&1 \ // RUN: | FileCheck -check-prefix FOOTER_PREPROC_GEN %s // FOOTER_PREPROC_GEN: clang{{.*}} "-fsycl-is-device"{{.*}} "-fsycl-int-header=[[INTHEADER:.+\.h]]" "-fsycl-int-footer=[[INTFOOTER:.+\h]]" "-sycl-std={{.*}}" "-o" "[[PREPROC_DEVICE:.+\.ii]]" // FOOTER_PREPROC_GEN: clang{{.*}} "-include" "[[INTHEADER]]"{{.*}} "-fsycl-is-host"{{.*}} "-E"{{.*}} "-o" "[[PREPROC1:.+\.ii]]" @@ -18,7 +18,7 @@ /// Preprocessed file use with integration footer // RUN: touch %t.ii -// RUN: %clangxx -fsycl -fsycl-use-footer %t.ii -### 2>&1 \ +// RUN: %clangxx -fsycl %t.ii -### 2>&1 \ // RUN: | FileCheck -check-prefix FOOTER_PREPROC_USE %s // FOOTER_PREPROC_USE: clang-offload-bundler{{.*}} "-outputs=[[HOST1:.+\.ii]],[[DEVICE_PP:.+\.ii]]" // FOOTER_PREPROC_USE: clang{{.*}} "-fsycl-is-device"{{.*}} "[[DEVICE_PP]]" diff --git a/clang/test/Driver/sycl-offload-win.c b/clang/test/Driver/sycl-offload-win.c index 1bfa24a1b5734..d89616e837923 100644 --- a/clang/test/Driver/sycl-offload-win.c +++ b/clang/test/Driver/sycl-offload-win.c @@ -101,5 +101,7 @@ // RUN: %clang_cl --target=x86_64-pc-windows-msvc -fsycl -P %s -### 2>&1 | FileCheck -check-prefix=FSYCL_P %s // FSYCL_P: clang{{.*}} "-cc1" "-triple" "spir64-unknown-unknown-sycldevice" {{.*}} "-E" {{.*}} "-o" "[[DEVICEPP:.+\.ii]]" // FSYCL_P: clang{{.*}} "-cc1" "-triple" "x86_64-pc-windows-msvc{{.*}}" {{.*}} "-E" {{.*}} "-o" "[[HOSTPP:.+\.ii]]" +// FSYCL_P: append-file{{.*}} "[[HOSTPP]]"{{.*}} "--output=[[APPEND:.+\.cpp]]" +// FSYCL_P: clang{{.*}} "-cc1" "-triple" "x86_64-pc-windows-msvc{{.*}}" {{.*}} "-E" {{.*}} "-o" "[[HOSTPP:.+\.ii]]"{{.*}} "[[APPEND]]" // FSYCL_P: clang-offload-bundler{{.*}} "-type=ii" "-targets=sycl-spir64-unknown-unknown-sycldevice,host-x86_64-pc-windows-msvc" {{.*}} "-inputs=[[DEVICEPP]],[[HOSTPP]]" diff --git a/clang/test/Driver/sycl-preprocess.cpp b/clang/test/Driver/sycl-preprocess.cpp index 0f9159cb0414e..beb8dd4b483fe 100644 --- a/clang/test/Driver/sycl-preprocess.cpp +++ b/clang/test/Driver/sycl-preprocess.cpp @@ -7,7 +7,9 @@ // RUN: | FileCheck -check-prefix PREPROC_ONLY %s // PREPROC_ONLY: clang{{.*}} "-fsycl-is-device"{{.*}} "-E"{{.*}} "-o" "[[DEVICE_OUT:.+\.ii]]" // PREPROC_ONLY: clang{{.*}} "-fsycl-is-device"{{.*}} "-fsycl-int-header=[[INTHEADER:.+\.h]]" "-fsycl-int-footer=[[INTFOOTER:.+\.h]]"{{.*}} "-fsyntax-only" -// PREPROC_ONLY: clang{{.*}} "-include" "[[INTHEADER]]"{{.*}} "-fsycl-is-host"{{.*}} "-o" "[[HOST_OUT:.+\.ii]]" +// PREPROC_ONLY: clang{{.*}} "-include" "[[INTHEADER]]"{{.*}} "-fsycl-is-host"{{.*}} "-o" "[[HOST_TEMP:.+\.ii]]" +// PREPROC_ONLY: append-file{{.*}} "[[HOST_TEMP]]" "--append=[[INTFOOTER]]" "--output=[[HOST_APPENDED:.+\.cpp]]" +// PREPROC_ONLY: clang{{.*}} "-fsycl-is-host"{{.*}} "-o" "[[HOST_OUT:.+\.ii]]"{{.*}} "[[HOST_APPENDED]]" // PREPROC_ONLY: clang-offload-bundler{{.*}} "-type=ii"{{.*}} "-outputs={{.+_output.ii}}" "-inputs=[[DEVICE_OUT]],[[HOST_OUT]]" /// When compiling from preprocessed file, no integration header is expected @@ -26,4 +28,6 @@ // PREPROC_PHASES: 4: compiler, {1}, none, (device-sycl) // PREPROC_PHASES: 5: offload, "host-sycl (x86_64-unknown-linux-gnu)" {3}, "device-sycl (spir64-unknown-unknown-sycldevice)" {4}, c++ // PREPROC_PHASES: 6: preprocessor, {5}, c++-cpp-output, (host-sycl) -// PREPROC_PHASES: 7: clang-offload-bundler, {2, 6}, c++-cpp-output, (host-sycl) +// PREPROC_PHASES: 7: append-footer, {6}, c++, (host-sycl) +// PREPROC_PHASES: 8: preprocessor, {7}, c++-cpp-output, (host-sycl) +// PREPROC_PHASES: 9: clang-offload-bundler, {2, 8}, c++-cpp-output, (host-sycl) From 7669de4d968bcecca1bbb563f671e68771a090f2 Mon Sep 17 00:00:00 2001 From: Alexey Sachkov Date: Wed, 16 Jun 2021 18:37:15 +0300 Subject: [PATCH 10/21] Add "-fsycl-device-only" to ESIMD diagnostics tests The reason for the change is that "-verify" mode does not work correctly with new integration footer approach, because it seems to ignore line markers in the input file. Since ESIMD tests for diagnostic messages only check device APIs and device compilation is the first step in our compilation flow, added "-fsycl-device-only" to avoid launching the host compilation. The positive side of this change is that now those tests should be a bit quicker as they do less work. The downside would be that we don't check for diagnostics coming out of host compilation. --- sycl/test/esimd/enums.cpp | 2 +- sycl/test/esimd/flat_atomic.cpp | 2 +- sycl/test/esimd/gather4_scatter4.cpp | 2 +- sycl/test/esimd/simd_copy_to_copy_from.cpp | 4 ++-- sycl/test/esimd/slm_atomic.cpp | 2 +- sycl/test/esimd/slm_load4.cpp | 2 +- 6 files changed, 7 insertions(+), 7 deletions(-) diff --git a/sycl/test/esimd/enums.cpp b/sycl/test/esimd/enums.cpp index 0f686d1038043..d91ea11579bb6 100644 --- a/sycl/test/esimd/enums.cpp +++ b/sycl/test/esimd/enums.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl -fsyntax-only -Xclang -verify %s +// RUN: %clangxx -fsycl -fsycl-device-only -fsyntax-only -Xclang -verify %s // This test checks compilation of various ESIMD enum types. Those which are // deprecated must produce deprecation messages. diff --git a/sycl/test/esimd/flat_atomic.cpp b/sycl/test/esimd/flat_atomic.cpp index 86a1800ca2158..22499835b72cd 100644 --- a/sycl/test/esimd/flat_atomic.cpp +++ b/sycl/test/esimd/flat_atomic.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl -fsyntax-only -Xclang -verify %s +// RUN: %clangxx -fsycl -fsycl-device-only -fsyntax-only -Xclang -verify %s // This test checks compilation of ESIMD atomic APIs. Those which are deprecated // must produce deprecation messages. diff --git a/sycl/test/esimd/gather4_scatter4.cpp b/sycl/test/esimd/gather4_scatter4.cpp index 31fb3ee241a4f..5e1ec851c0046 100644 --- a/sycl/test/esimd/gather4_scatter4.cpp +++ b/sycl/test/esimd/gather4_scatter4.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl -fsyntax-only -Xclang -verify %s +// RUN: %clangxx -fsycl -fsycl-device-only -fsyntax-only -Xclang -verify %s // This test checks compilation of ESIMD slm gather4/scatter4 APIs. Those which // are deprecated must produce deprecation messages. diff --git a/sycl/test/esimd/simd_copy_to_copy_from.cpp b/sycl/test/esimd/simd_copy_to_copy_from.cpp index 360c5c3fdc28a..40f077da3631c 100644 --- a/sycl/test/esimd/simd_copy_to_copy_from.cpp +++ b/sycl/test/esimd/simd_copy_to_copy_from.cpp @@ -1,6 +1,6 @@ -// RUN: %clangxx -fsycl -fsyntax-only -Xclang -verify %s +// RUN: %clangxx -fsycl -fsycl-device-only -fsyntax-only -Xclang -verify %s -// This test checks that both host and device compilers can: +// This test checks that device compiler can: // - successfully compile simd::copy_to and simd::copy_from APIs // - emit an error if argument of an incompatible type is used // in place of the accessor argument diff --git a/sycl/test/esimd/slm_atomic.cpp b/sycl/test/esimd/slm_atomic.cpp index fce324854a6e4..e74ddc2cda598 100644 --- a/sycl/test/esimd/slm_atomic.cpp +++ b/sycl/test/esimd/slm_atomic.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl -fsyntax-only -Xclang -verify %s +// RUN: %clangxx -fsycl -fsycl-device-only -fsyntax-only -Xclang -verify %s // This test checks compilation of ESIMD slm atomic APIs. Those which are // deprecated must produce deprecation messages. diff --git a/sycl/test/esimd/slm_load4.cpp b/sycl/test/esimd/slm_load4.cpp index 163c06010c89c..fadaa0ced6e72 100644 --- a/sycl/test/esimd/slm_load4.cpp +++ b/sycl/test/esimd/slm_load4.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl -fsyntax-only -Xclang -verify %s +// RUN: %clangxx -fsycl -fsycl-device-only -fsyntax-only -Xclang -verify %s // This test checks compilation of ESIMD slm load4/store4 APIs. Those which are // deprecated must produce deprecation messages. From d86c37ae1d8a17aabdd7d0c11c9b07aac3ebc831 Mon Sep 17 00:00:00 2001 From: Alexey Sachkov Date: Wed, 16 Jun 2021 23:04:51 +0300 Subject: [PATCH 11/21] temp --- sycl/test/CMakeLists.txt | 2 +- sycl/test/basic_tests/range_error.cpp | 15 +++++---------- 2 files changed, 6 insertions(+), 11 deletions(-) diff --git a/sycl/test/CMakeLists.txt b/sycl/test/CMakeLists.txt index a2a12c8797a76..d8b8e02ae7d63 100644 --- a/sycl/test/CMakeLists.txt +++ b/sycl/test/CMakeLists.txt @@ -2,7 +2,7 @@ set(LLVM_TOOLS_DIR "${LLVM_BINARY_DIR}/bin/") get_target_property(SYCL_BINARY_DIR sycl-toolchain BINARY_DIR) -set(SYCL_INCLUDE "${SYCL_INCLUDE_BUILD_DIR}") +set(SYCL_INCLUDE "${SYCL_INCLUDE_BUILD_DIR}/sycl/") set(SYCL_TOOLS_SRC_DIR "${PROJECT_SOURCE_DIR}/tools/") set(LLVM_BUILD_BINARY_DIRS "${LLVM_BINARY_DIR}/bin/") set(LLVM_BUILD_LIBRARY_DIRS "${LLVM_BINARY_DIR}/lib/") diff --git a/sycl/test/basic_tests/range_error.cpp b/sycl/test/basic_tests/range_error.cpp index c11d2242119c4..7a9251d2b9376 100644 --- a/sycl/test/basic_tests/range_error.cpp +++ b/sycl/test/basic_tests/range_error.cpp @@ -1,11 +1,4 @@ -// RUN: %clangxx -fsycl -Xclang -verify %s -Xclang -verify-ignore-unexpected=note,warning -fsyntax-only -//==--------------- range_error.cpp - SYCL range error test ----------------==// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===----------------------------------------------------------------------===// +// RUN: %clangxx %s -std=c++17 -fsyntax-only -Xclang -fsycl-is-host -Xclang -verify -Xclang -verify-ignore-unexpected=note,warning -isystem %sycl_include -isystem %opencl_include_dir #include #include #include @@ -33,6 +26,8 @@ int main() { assert(three_dim_range.get(2) ==2); assert(three_dim_range[2] ==2); cout << "three_dim_range passed " << endl; - cl::sycl::range<1> one_dim_range_f1(64, 2, 4);//expected-error {{no matching constructor for initialization of 'cl::sycl::range<1>'}} - cl::sycl::range<2> two_dim_range_f1(64);//expected-error {{no matching constructor for initialization of 'cl::sycl::range<2>'}} + // expected-error@+1 {{no matching constructor for initialization of 'cl::sycl::range<1>'}} + cl::sycl::range<1> one_dim_range_f1(64, 2, 4); + // expected-error@+1 {{no matching constructor for initialization of 'cl::sycl::range<2>'}} + cl::sycl::range<2> two_dim_range_f1(64); } From 33f1ed9acef5cb1b419747843476d2cb50fad7b3 Mon Sep 17 00:00:00 2001 From: Alexey Sachkov Date: Fri, 18 Jun 2021 18:22:42 +0300 Subject: [PATCH 12/21] Fix host-only tests for diagnostics Corrected '%sycl_include' substitution. Added '%fsycl-host-only' substituion, which contains a mininum amount of options which are needed to launch host-only compilation in SYCL mode, which is useful for "-verify" tests --- sycl/test/basic_tests/image_accessor_types.cpp | 8 +++----- sycl/test/basic_tests/implicit_conversion_error.cpp | 10 ++-------- sycl/test/basic_tests/range_error.cpp | 2 +- sycl/test/basic_tests/set_arg_error.cpp | 2 +- sycl/test/basic_tests/vectors/ctad_fail.cpp | 9 +-------- sycl/test/gdb/accessors.cpp | 2 +- sycl/test/gdb/printers.cpp | 2 +- sycl/test/lit.cfg.py | 2 ++ sycl/test/separate-compile/test.cpp | 4 ++-- sycl/test/warnings/sycl_2020_deprecations.cpp | 6 ++---- 10 files changed, 16 insertions(+), 31 deletions(-) diff --git a/sycl/test/basic_tests/image_accessor_types.cpp b/sycl/test/basic_tests/image_accessor_types.cpp index 5bcdc22357295..6abbca5851896 100644 --- a/sycl/test/basic_tests/image_accessor_types.cpp +++ b/sycl/test/basic_tests/image_accessor_types.cpp @@ -1,4 +1,4 @@ -// RUN: not %clangxx -fsyntax-only -std=c++17 %s -I %sycl_include/sycl 2>&1 | FileCheck %s +// RUN: %clangxx -fsyntax-only %fsycl-host-only -Xclang -verify -Xclang -verify-ignore-unexpected=note %s #include #include @@ -12,12 +12,10 @@ int main() { myQueue.submit([&](handler &cgh) { accessor NotValidType1( srcImage, cgh); - // CHECK: The data type of an image accessor must be only cl_int4, cl_uint4, - // CHECK-SAME: cl_float4 or cl_half4 + // expected-error@CL/sycl/accessor.hpp:* {{The data type of an image accessor must be only cl_int4, cl_uint4, cl_float4 or cl_half4}} accessor NotValidType2( srcImage, cgh); - // CHECK: The data type of an image accessor must be only cl_int4, cl_uint4, - // CHECK-SAME: cl_float4 or cl_half4 + // expected-error@CL/sycl/accessor.hpp:* {{The data type of an image accessor must be only cl_int4, cl_uint4, cl_float4 or cl_half4}} accessor ValidSYCLFloat(srcImage, cgh); accessor ValidSYCLInt( diff --git a/sycl/test/basic_tests/implicit_conversion_error.cpp b/sycl/test/basic_tests/implicit_conversion_error.cpp index 6b4b389c45f70..2cf5325a08d48 100644 --- a/sycl/test/basic_tests/implicit_conversion_error.cpp +++ b/sycl/test/basic_tests/implicit_conversion_error.cpp @@ -1,11 +1,5 @@ -// RUN: %clangxx -fsycl -fsyntax-only -Xclang -verify %s -Xclang -verify-ignore-unexpected=note,warning -//=- implicit_conversion_error.cpp - Unintended implicit conversion check -=// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===-----------------------------------------------------------------------===// +// RUN: %clangxx -fsyntax-only %fsycl-host-only -Xclang -verify -Xclang -verify-ignore-unexpected=note,warning %s + #include int main() { diff --git a/sycl/test/basic_tests/range_error.cpp b/sycl/test/basic_tests/range_error.cpp index 7a9251d2b9376..2aed44d9eb7ca 100644 --- a/sycl/test/basic_tests/range_error.cpp +++ b/sycl/test/basic_tests/range_error.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx %s -std=c++17 -fsyntax-only -Xclang -fsycl-is-host -Xclang -verify -Xclang -verify-ignore-unexpected=note,warning -isystem %sycl_include -isystem %opencl_include_dir +// RUN: %clangxx %s %fsycl-host-only -fsyntax-only -Xclang -verify -Xclang -verify-ignore-unexpected=note,warning #include #include #include diff --git a/sycl/test/basic_tests/set_arg_error.cpp b/sycl/test/basic_tests/set_arg_error.cpp index d4215766c7a56..c7ea5be42808e 100644 --- a/sycl/test/basic_tests/set_arg_error.cpp +++ b/sycl/test/basic_tests/set_arg_error.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl -Xclang -verify %s -I %sycl_include -Xclang -verify-ignore-unexpected=note,warning -fsyntax-only +// RUN: %clangxx %fsycl-host-only -fsyntax-only -Xclang -verify -Xclang -verify-ignore-unexpected=note,warning %s #include diff --git a/sycl/test/basic_tests/vectors/ctad_fail.cpp b/sycl/test/basic_tests/vectors/ctad_fail.cpp index bdae6b013e62c..b584cebb698d0 100644 --- a/sycl/test/basic_tests/vectors/ctad_fail.cpp +++ b/sycl/test/basic_tests/vectors/ctad_fail.cpp @@ -1,11 +1,4 @@ -// RUN: %clangxx -std=c++17 -fsyntax-only -Xclang -verify %s -I %sycl_include/sycl -Xclang -verify-ignore-unexpected -//==--------------- ctad.cpp - SYCL vector CTAD fail test ------------------==// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===----------------------------------------------------------------------===// +// RUN: %clangxx %fsycl-host-only -fsyntax-only -Xclang -verify -Xclang -verify-ignore-unexpected %s #include namespace sycl = cl::sycl; diff --git a/sycl/test/gdb/accessors.cpp b/sycl/test/gdb/accessors.cpp index da7cc84ecb2ea..adc8fb369a222 100644 --- a/sycl/test/gdb/accessors.cpp +++ b/sycl/test/gdb/accessors.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -c -fno-color-diagnostics -std=c++17 -I %sycl_include/sycl -Xclang -ast-dump %s | FileCheck %s +// RUN: %clangxx -c -fno-color-diagnostics -std=c++17 -I %sycl_include -Xclang -ast-dump %s | FileCheck %s // UNSUPPORTED: windows #include diff --git a/sycl/test/gdb/printers.cpp b/sycl/test/gdb/printers.cpp index 58fda0db57034..fe81f9892cf9c 100644 --- a/sycl/test/gdb/printers.cpp +++ b/sycl/test/gdb/printers.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -c -fno-color-diagnostics -std=c++17 -I %sycl_include/sycl -Xclang -ast-dump %s | FileCheck %s +// RUN: %clangxx -c -fno-color-diagnostics -std=c++17 -I %sycl_include -Xclang -ast-dump %s | FileCheck %s // UNSUPPORTED: windows #include #include diff --git a/sycl/test/lit.cfg.py b/sycl/test/lit.cfg.py index b41a369aa90f6..2131d64ffabd4 100644 --- a/sycl/test/lit.cfg.py +++ b/sycl/test/lit.cfg.py @@ -90,6 +90,8 @@ config.substitutions.append( ('%llvm_build_lib_dir', config.llvm_build_lib_dir ) ) config.substitutions.append( ('%llvm_build_bin_dir', config.llvm_build_bin_dir ) ) +config.substitutions.append( ('%fsycl-host-only', '-std=c++17 -Xclang -fsycl-is-host -isystem %s -isystem %s' % (config.sycl_include, config.opencl_include_dir) ) ) + llvm_config.add_tool_substitutions(['llvm-spirv'], [config.sycl_tools_dir]) config.substitutions.append( ('%RUN_ON_HOST', "env SYCL_DEVICE_FILTER=host ") ) diff --git a/sycl/test/separate-compile/test.cpp b/sycl/test/separate-compile/test.cpp index df97ae177d5d1..30e58f50aee9b 100644 --- a/sycl/test/separate-compile/test.cpp +++ b/sycl/test/separate-compile/test.cpp @@ -7,13 +7,13 @@ // >> host compilation... // Driver automatically adds -D_DLL and -D_MT on Windows with -fsycl. // Add those defines required for Windows and harmless for Linux. -// RUN: %clangxx -D_DLL -D_MT -std=c++17 -include sycl_ihdr_a.h -g -c %s -o a.o -I %sycl_include/sycl -Wno-sycl-strict +// RUN: %clangxx -D_DLL -D_MT -std=c++17 -include sycl_ihdr_a.h -g -c %s -o a.o -I %sycl_include -Wno-sycl-strict // // >> ---- compile src2 // >> device compilation... // RUN: %clangxx -DB_CPP=1 -fsycl-device-only -Xclang -fsycl-int-header=sycl_ihdr_b.h %s -c -o b_kernel.bc -Wno-sycl-strict // >> host compilation... -// RUN: %clangxx -DB_CPP=1 -D_DLL -D_MT -std=c++17 -include sycl_ihdr_b.h -g -c %s -o b.o -I %sycl_include/sycl -Wno-sycl-strict +// RUN: %clangxx -DB_CPP=1 -D_DLL -D_MT -std=c++17 -include sycl_ihdr_b.h -g -c %s -o b.o -I %sycl_include -Wno-sycl-strict // // >> ---- bundle .o with .spv // >> run bundler diff --git a/sycl/test/warnings/sycl_2020_deprecations.cpp b/sycl/test/warnings/sycl_2020_deprecations.cpp index 5e159afa496cd..023b4b9d9f905 100644 --- a/sycl/test/warnings/sycl_2020_deprecations.cpp +++ b/sycl/test/warnings/sycl_2020_deprecations.cpp @@ -1,7 +1,5 @@ -// RUN: %clangxx -fsycl -sycl-std=2020 -Xclang -verify -Xclang -verify-ignore-unexpected=note %s -o %t.out -// RUN: %clangxx -fsycl -Xclang -verify -Xclang -verify-ignore-unexpected=note %s -o %t.out -// RUN: %clangxx -fsycl -sycl-std=2017 -Werror %s -o %t.out -// RUN: %clangxx -fsycl -sycl-std=1.2.1 -Werror %s -o %t.out +// RUN: %clangxx %fsycl-host-only -fsyntax-only -sycl-std=2020 -Xclang -verify -Xclang -verify-ignore-unexpected=note %s -o %t.out +// RUN: %clangxx %fsycl-host-only -fsyntax-only -Xclang -verify -Xclang -verify-ignore-unexpected=note %s -o %t.out #include From b3a70768bc48351ea27f5e3e15e611be3b200c31 Mon Sep 17 00:00:00 2001 From: Alexey Sachkov Date: Fri, 18 Jun 2021 18:50:40 +0300 Subject: [PATCH 13/21] Fix one more ESIMD test after merge from upstream --- sycl/test/esimd/simd_subscript.cpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/sycl/test/esimd/simd_subscript.cpp b/sycl/test/esimd/simd_subscript.cpp index 35f90c231377f..8bd24fc182b74 100644 --- a/sycl/test/esimd/simd_subscript.cpp +++ b/sycl/test/esimd/simd_subscript.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl -fsyntax-only -Xclang -verify %s +// RUN: %clangxx -fsycl -fsycl-device-only -fsyntax-only -Xclang -verify %s #include From a51376975899daf9861fa94ab8dbea9268deea3f Mon Sep 17 00:00:00 2001 From: Alexey Sachkov Date: Mon, 21 Jun 2021 15:28:05 +0300 Subject: [PATCH 14/21] temp --- sycl/test/CMakeLists.txt | 2 +- .../test/basic_tests/image_accessor_types.cpp | 1 + sycl/test/esimd/enums.cpp | 33 +++++++++++++------ sycl/test/esimd/flat_atomic.cpp | 23 +++++++++---- sycl/test/esimd/gather4_scatter4.cpp | 29 +++++++++++----- sycl/test/esimd/simd_copy_to_copy_from.cpp | 21 ++++-------- sycl/test/esimd/simd_subscript.cpp | 3 +- sycl/test/esimd/slm_atomic.cpp | 23 +++++++++---- sycl/test/esimd/slm_load4.cpp | 17 +++++++--- sycl/test/gdb/accessors.cpp | 2 +- sycl/test/gdb/printers.cpp | 2 +- sycl/test/lit.cfg.py | 2 +- sycl/test/separate-compile/test.cpp | 4 +-- sycl/test/warnings/sycl_2020_deprecations.cpp | 2 ++ 14 files changed, 105 insertions(+), 59 deletions(-) diff --git a/sycl/test/CMakeLists.txt b/sycl/test/CMakeLists.txt index d8b8e02ae7d63..a2a12c8797a76 100644 --- a/sycl/test/CMakeLists.txt +++ b/sycl/test/CMakeLists.txt @@ -2,7 +2,7 @@ set(LLVM_TOOLS_DIR "${LLVM_BINARY_DIR}/bin/") get_target_property(SYCL_BINARY_DIR sycl-toolchain BINARY_DIR) -set(SYCL_INCLUDE "${SYCL_INCLUDE_BUILD_DIR}/sycl/") +set(SYCL_INCLUDE "${SYCL_INCLUDE_BUILD_DIR}") set(SYCL_TOOLS_SRC_DIR "${PROJECT_SOURCE_DIR}/tools/") set(LLVM_BUILD_BINARY_DIRS "${LLVM_BINARY_DIR}/bin/") set(LLVM_BUILD_LIBRARY_DIRS "${LLVM_BINARY_DIR}/lib/") diff --git a/sycl/test/basic_tests/image_accessor_types.cpp b/sycl/test/basic_tests/image_accessor_types.cpp index 6abbca5851896..4133d75954444 100644 --- a/sycl/test/basic_tests/image_accessor_types.cpp +++ b/sycl/test/basic_tests/image_accessor_types.cpp @@ -1,4 +1,5 @@ // RUN: %clangxx -fsyntax-only %fsycl-host-only -Xclang -verify -Xclang -verify-ignore-unexpected=note %s +// RUN: %clangxx -fsyntax-only -fsycl -fsycl-device-only -Xclang -verify -Xclang -verify-ignore-unexpected=note %s #include #include diff --git a/sycl/test/esimd/enums.cpp b/sycl/test/esimd/enums.cpp index d91ea11579bb6..d0a5b57dcfd8e 100644 --- a/sycl/test/esimd/enums.cpp +++ b/sycl/test/esimd/enums.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl -fsycl-device-only -fsyntax-only -Xclang -verify %s +// RUN: %clangxx -fsycl -fsyntax-only %s 2>&1 | FileCheck %s // This test checks compilation of various ESIMD enum types. Those which are // deprecated must produce deprecation messages. @@ -8,21 +8,34 @@ using namespace sycl::ext::intel::experimental::esimd; void foo() SYCL_ESIMD_FUNCTION { - // These should produce deprecation messages: + // These should produce deprecation messages for both device: int x; - // expected-warning@+2 {{deprecated}} - // expected-note@sycl/ext/intel/experimental/esimd/common.hpp:* {{}} + // CHECK: enums.cpp:15{{.*}}warning: 'WAIT' is deprecated + // CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: x = static_cast(ESIMD_SBARRIER_WAIT); - // expected-warning@+2 {{deprecated}} - // expected-note@sycl/ext/intel/experimental/esimd/common.hpp:* {{}} + // CHECK: enums.cpp:18{{.*}}warning: 'ATOMIC_ADD' is deprecated + // CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: x = static_cast(EsimdAtomicOpType::ATOMIC_ADD); - // expected-warning@+2 {{deprecated}} - // expected-note@sycl/ext/intel/experimental/esimd/common.hpp:* {{}} + // CHECK: enums.cpp:21{{.*}}warning: 'ESIMD_R_ENABLE' is deprecated + // CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: x = static_cast(ChannelMaskType::ESIMD_R_ENABLE); - // expected-warning@+2 {{deprecated}} - // expected-note@sycl/ext/intel/experimental/esimd/common.hpp:* {{}} + // CHECK: enums.cpp:24{{.*}}warning: 'GENX_NOSAT' is deprecated + // CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: x = static_cast(GENX_NOSAT); + // A "border" between host and device compilations + // CHECK-LABEL: 4 warnings generated + + // And for host: + // CHECK: enums.cpp:15{{.*}}warning: 'WAIT' is deprecated + // CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: + // CHECK: enums.cpp:18{{.*}}warning: 'ATOMIC_ADD' is deprecated + // CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: + // CHECK: enums.cpp:21{{.*}}warning: 'ESIMD_R_ENABLE' is deprecated + // CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: + // CHECK: enums.cpp:24{{.*}}warning: 'GENX_NOSAT' is deprecated + // CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: + // These should compile cleanly: x = static_cast(split_barrier_action::wait); x = static_cast(atomic_op::add); diff --git a/sycl/test/esimd/flat_atomic.cpp b/sycl/test/esimd/flat_atomic.cpp index 22499835b72cd..5a749789b5e9d 100644 --- a/sycl/test/esimd/flat_atomic.cpp +++ b/sycl/test/esimd/flat_atomic.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl -fsycl-device-only -fsyntax-only -Xclang -verify %s +// RUN: %clangxx -fsycl -fsyntax-only %s 2>&1 | FileCheck %s // This test checks compilation of ESIMD atomic APIs. Those which are deprecated // must produce deprecation messages. @@ -15,8 +15,8 @@ void kernel0(accessor &buf) SYCL_ESIMD_FUNCTION { simd offsets(0, 1); - // expected-warning@+2 {{deprecated}} - // expected-note@sycl/ext/intel/experimental/esimd/common.hpp:* {{}} + // CHECK: flat_atomic.cpp:20{{.*}}warning: 'ATOMIC_INC' is deprecated + // CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: flat_atomic(buf.get_pointer(), offsets, 1); flat_atomic(buf.get_pointer(), offsets, 1); } @@ -26,8 +26,8 @@ void kernel1(accessor offsets(0, 1); simd v1(0, 1); - // expected-warning@+2 {{deprecated}} - // expected-note@sycl/ext/intel/experimental/esimd/common.hpp:* {{}} + // CHECK: flat_atomic.cpp:31{{.*}}warning: 'ATOMIC_ADD' is deprecated + // CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: flat_atomic(buf.get_pointer(), offsets, v1, 1); flat_atomic(buf.get_pointer(), offsets, v1, 1); } @@ -37,9 +37,18 @@ void kernel2(accessor offsets(0, 1); simd v1(0, 1); - // expected-warning@+2 {{deprecated}} - // expected-note@sycl/ext/intel/experimental/esimd/common.hpp:* {{}} + // CHECK: flat_atomic.cpp:42{{.*}}warning: 'ATOMIC_CMPXCHG' is deprecated + // CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: flat_atomic(buf.get_pointer(), offsets, v1, v1, 1); flat_atomic(buf.get_pointer(), offsets, v1, v1, 1); } + +// A "border" between device and host compilations +// CHECK-LABEL: 3 warnings generated +// CHECK: flat_atomic.cpp:20{{.*}}warning: 'ATOMIC_INC' is deprecated +// CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: +// CHECK: flat_atomic.cpp:31{{.*}}warning: 'ATOMIC_ADD' is deprecated +// CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: +// CHECK: flat_atomic.cpp:42{{.*}}warning: 'ATOMIC_CMPXCHG' is deprecated +// CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: diff --git a/sycl/test/esimd/gather4_scatter4.cpp b/sycl/test/esimd/gather4_scatter4.cpp index 5e1ec851c0046..5328a6faed74a 100644 --- a/sycl/test/esimd/gather4_scatter4.cpp +++ b/sycl/test/esimd/gather4_scatter4.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl -fsycl-device-only -fsyntax-only -Xclang -verify %s +// RUN: %clangxx -fsycl -fsyntax-only %s 2>&1 | FileCheck %s // This test checks compilation of ESIMD slm gather4/scatter4 APIs. Those which // are deprecated must produce deprecation messages. @@ -16,23 +16,34 @@ void kernel(accessor offsets(0, 1); simd v1(0, 1); - // expected-warning@+2 {{deprecated}} - // expected-note@sycl/ext/intel/experimental/esimd/common.hpp:* {{}} + // CHECK: gather4_scatter4.cpp:21{{.*}}warning: 'ESIMD_ABGR_ENABLE' is deprecated + // CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: auto v0 = gather4(buf.get_pointer(), offsets); - // expected-warning@+2 {{deprecated}} - // expected-note@sycl/ext/intel/experimental/esimd/common.hpp:* {{}} + // CHECK: gather4_scatter4.cpp:24{{.*}}warning: 'ESIMD_ABGR_ENABLE' is deprecated + // CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: v0 = gather4(buf.get_pointer(), offsets); v0 = gather4(buf.get_pointer(), offsets); v0 = v0 + v1; - // expected-warning@+2 {{deprecated}} - // expected-note@sycl/ext/intel/experimental/esimd/common.hpp:* {{}} + // CHECK: gather4_scatter4.cpp:32{{.*}}warning: 'ESIMD_ABGR_ENABLE' is deprecated + // CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: scatter4(buf.get_pointer(), v0, offsets); - // expected-warning@+2 {{deprecated}} - // expected-note@sycl/ext/intel/experimental/esimd/common.hpp:* {{}} + // CHECK: gather4_scatter4.cpp:35{{.*}}warning: 'ESIMD_ABGR_ENABLE' is deprecated + // CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: scatter4(buf.get_pointer(), v0, offsets); scatter4(buf.get_pointer(), v0, offsets); } + +// A "border" between host and device compilations +// CHECK-LABEL: 4 warnings generated +// CHECK: gather4_scatter4.cpp:21{{.*}}warning: 'ESIMD_ABGR_ENABLE' is deprecated +// CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: +// CHECK: gather4_scatter4.cpp:24{{.*}}warning: 'ESIMD_ABGR_ENABLE' is deprecated +// CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: +// CHECK: gather4_scatter4.cpp:32{{.*}}warning: 'ESIMD_ABGR_ENABLE' is deprecated +// CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: +// CHECK: gather4_scatter4.cpp:35{{.*}}warning: 'ESIMD_ABGR_ENABLE' is deprecated +// CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: diff --git a/sycl/test/esimd/simd_copy_to_copy_from.cpp b/sycl/test/esimd/simd_copy_to_copy_from.cpp index 40f077da3631c..13f4e05c5c794 100644 --- a/sycl/test/esimd/simd_copy_to_copy_from.cpp +++ b/sycl/test/esimd/simd_copy_to_copy_from.cpp @@ -1,6 +1,7 @@ -// RUN: %clangxx -fsycl -fsycl-device-only -fsyntax-only -Xclang -verify %s +// RUN: not %clangxx -fsycl -fsycl-device-only -fsyntax-only %s 2>&1 | FileCheck %s +// RUN: not %clangxx %fsycl-host-only -fsyntax-only %s 2>&1 | FileCheck %s -// This test checks that device compiler can: +// This test checks that device and host compilers can: // - successfully compile simd::copy_to and simd::copy_from APIs // - emit an error if argument of an incompatible type is used // in place of the accessor argument @@ -41,14 +42,10 @@ kernel3(accessor &buf) SYCL_ESIMD_FUNCTION { simd v1(0, 1); simd v0; - // expected-error@+3 {{no matching member function for call to 'copy_from'}} - // expected-note@sycl/ext/intel/experimental/esimd/simd.hpp:* {{}} - // expected-note@sycl/ext/intel/experimental/esimd/simd.hpp:* {{}} + // CHECK: simd_copy_to_copy_from.cpp:46{{.*}}error: no matching member function for call to 'copy_from' v0.copy_from(buf, 0); v0 = v0 + v1; - // expected-error@+3 {{no matching member function for call to 'copy_to'}} - // expected-note@sycl/ext/intel/experimental/esimd/simd.hpp:* {{}} - // expected-note@sycl/ext/intel/experimental/esimd/simd.hpp:* {{}} + // CHECK: simd_copy_to_copy_from.cpp:49{{.*}}error: no matching member function for call to 'copy_to' v0.copy_to(buf, 0); } @@ -57,9 +54,7 @@ SYCL_EXTERNAL void kernel4( accessor &buf) SYCL_ESIMD_FUNCTION { simd v; - // expected-error@+3 {{no matching member function for call to 'copy_from'}} - // expected-note@sycl/ext/intel/experimental/esimd/simd.hpp:* {{}} - // expected-note@sycl/ext/intel/experimental/esimd/simd.hpp:* {{}} + // CHECK: simd_copy_to_copy_from.cpp:58{{.*}}error: no matching member function for call to 'copy_from' v.copy_from(buf, 0); } @@ -68,8 +63,6 @@ SYCL_EXTERNAL void kernel5( accessor &buf) SYCL_ESIMD_FUNCTION { simd v(0, 1); - // expected-error@+3 {{no matching member function for call to 'copy_to'}} - // expected-note@sycl/ext/intel/experimental/esimd/simd.hpp:* {{}} - // expected-note@sycl/ext/intel/experimental/esimd/simd.hpp:* {{}} + // CHECK: simd_copy_to_copy_from.cpp:67{{.*}}error: no matching member function for call to 'copy_to' v.copy_to(buf, 0); } diff --git a/sycl/test/esimd/simd_subscript.cpp b/sycl/test/esimd/simd_subscript.cpp index 8bd24fc182b74..6f5139388235f 100644 --- a/sycl/test/esimd/simd_subscript.cpp +++ b/sycl/test/esimd/simd_subscript.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl -fsycl-device-only -fsyntax-only -Xclang -verify %s +// RUN: not %clangxx -fsycl -fsyntax-only %s 2>&1 | FileCheck %s #include @@ -68,6 +68,7 @@ void test_simd_writable_subscript() SYCL_ESIMD_FUNCTION { void test_simd_const_subscript() SYCL_ESIMD_FUNCTION { const simd cv = 1; int val2 = cv[0]; // returns int instead of simd_view + // CHECK: simd_subscript.cpp:72{{.*}}error: expression is not assignable cv[1] = 0; // expected-error {{expression is not assignable}} } diff --git a/sycl/test/esimd/slm_atomic.cpp b/sycl/test/esimd/slm_atomic.cpp index e74ddc2cda598..b8a0238e57633 100644 --- a/sycl/test/esimd/slm_atomic.cpp +++ b/sycl/test/esimd/slm_atomic.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl -fsycl-device-only -fsyntax-only -Xclang -verify %s +// RUN: %clangxx -fsycl -fsyntax-only %s 2>&1 | FileCheck %s // This test checks compilation of ESIMD slm atomic APIs. Those which are // deprecated must produce deprecation messages. @@ -14,8 +14,8 @@ using namespace cl::sycl; void kernel0() SYCL_ESIMD_FUNCTION { simd offsets(0, 1); - // expected-warning@+2 {{deprecated}} - // expected-note@sycl/ext/intel/experimental/esimd/common.hpp:* {{}} + // CHECK: slm_atomic.cpp:19{{.*}}warning: 'ATOMIC_INC' is deprecated + // CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: slm_atomic(offsets, 1); slm_atomic(offsets, 1); } @@ -24,8 +24,8 @@ void kernel1() SYCL_ESIMD_FUNCTION { simd offsets(0, 1); simd v1(0, 1); - // expected-warning@+2 {{deprecated}} - // expected-note@sycl/ext/intel/experimental/esimd/common.hpp:* {{}} + // CHECK: slm_atomic.cpp:29{{.*}}warning: 'ATOMIC_ADD' is deprecated + // CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: slm_atomic(offsets, v1, 1); slm_atomic(offsets, v1, 1); } @@ -34,8 +34,17 @@ void kernel2() SYCL_ESIMD_FUNCTION { simd offsets(0, 1); simd v1(0, 1); - // expected-warning@+2 {{deprecated}} - // expected-note@sycl/ext/intel/experimental/esimd/common.hpp:* {{}} + // CHECK: slm_atomic.cpp:39{{.*}}warning: 'ATOMIC_CMPXCHG' is deprecated + // CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: slm_atomic(offsets, v1, v1, 1); slm_atomic(offsets, v1, v1, 1); } + +// A "border" between device and host compilations +// CHECK-LABEL: 3 warnings generated +// CHECK: slm_atomic.cpp:19{{.*}}warning: 'ATOMIC_INC' is deprecated +// CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: +// CHECK: slm_atomic.cpp:29{{.*}}warning: 'ATOMIC_ADD' is deprecated +// CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: +// CHECK: slm_atomic.cpp:39{{.*}}warning: 'ATOMIC_CMPXCHG' is deprecated +// CHECK: sycl/ext/intel/experimental/esimd/common.hpp{{.*}}note: diff --git a/sycl/test/esimd/slm_load4.cpp b/sycl/test/esimd/slm_load4.cpp index fadaa0ced6e72..4684dff34be10 100644 --- a/sycl/test/esimd/slm_load4.cpp +++ b/sycl/test/esimd/slm_load4.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -fsycl -fsycl-device-only -fsyntax-only -Xclang -verify %s +// RUN: %clangxx -fsycl -fsyntax-only %s 2>&1 | FileCheck %s // This test checks compilation of ESIMD slm load4/store4 APIs. Those which are // deprecated must produce deprecation messages. @@ -14,15 +14,22 @@ void caller() SYCL_ESIMD_FUNCTION { slm_init(1024); - // expected-warning@+2 {{deprecated}} - // expected-note@sycl/ext/intel/experimental/esimd/common.hpp:* {{}} + // CHECK: slm_load4.cpp:19{{.*}}warning: 'ESIMD_ABGR_ENABLE' is deprecated + // CHECK: sycl/ext/intel/experimental/esimd/common.hpp:{{.*}}note: auto v0 = slm_load4(offsets); v0 = slm_load4(offsets); v0 = v0 + v1; - // expected-warning@+2 {{deprecated}} - // expected-note@sycl/ext/intel/experimental/esimd/common.hpp:* {{}} + // CHECK: slm_load4.cpp:26{{.*}}warning: 'ESIMD_ABGR_ENABLE' is deprecated + // CHECK: sycl/ext/intel/experimental/esimd/common.hpp:{{.*}}note: slm_store4(v0, offsets); slm_store4(v0, offsets); } + +// A "border" between host and device compilations +// CHECK-LABEL: 2 warnings generated +// CHECK: slm_load4.cpp:19{{.*}}warning: 'ESIMD_ABGR_ENABLE' is deprecated +// CHECK: sycl/ext/intel/experimental/esimd/common.hpp:{{.*}}note: +// CHECK: slm_load4.cpp:26{{.*}}warning: 'ESIMD_ABGR_ENABLE' is deprecated +// CHECK: sycl/ext/intel/experimental/esimd/common.hpp:{{.*}}note: diff --git a/sycl/test/gdb/accessors.cpp b/sycl/test/gdb/accessors.cpp index adc8fb369a222..da7cc84ecb2ea 100644 --- a/sycl/test/gdb/accessors.cpp +++ b/sycl/test/gdb/accessors.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -c -fno-color-diagnostics -std=c++17 -I %sycl_include -Xclang -ast-dump %s | FileCheck %s +// RUN: %clangxx -c -fno-color-diagnostics -std=c++17 -I %sycl_include/sycl -Xclang -ast-dump %s | FileCheck %s // UNSUPPORTED: windows #include diff --git a/sycl/test/gdb/printers.cpp b/sycl/test/gdb/printers.cpp index fe81f9892cf9c..58fda0db57034 100644 --- a/sycl/test/gdb/printers.cpp +++ b/sycl/test/gdb/printers.cpp @@ -1,4 +1,4 @@ -// RUN: %clangxx -c -fno-color-diagnostics -std=c++17 -I %sycl_include -Xclang -ast-dump %s | FileCheck %s +// RUN: %clangxx -c -fno-color-diagnostics -std=c++17 -I %sycl_include/sycl -Xclang -ast-dump %s | FileCheck %s // UNSUPPORTED: windows #include #include diff --git a/sycl/test/lit.cfg.py b/sycl/test/lit.cfg.py index 2131d64ffabd4..70b09741b2df5 100644 --- a/sycl/test/lit.cfg.py +++ b/sycl/test/lit.cfg.py @@ -90,7 +90,7 @@ config.substitutions.append( ('%llvm_build_lib_dir', config.llvm_build_lib_dir ) ) config.substitutions.append( ('%llvm_build_bin_dir', config.llvm_build_bin_dir ) ) -config.substitutions.append( ('%fsycl-host-only', '-std=c++17 -Xclang -fsycl-is-host -isystem %s -isystem %s' % (config.sycl_include, config.opencl_include_dir) ) ) +config.substitutions.append( ('%fsycl-host-only', '-std=c++17 -Xclang -fsycl-is-host -isystem %s -isystem %s -isystem %s' % (config.sycl_include, config.opencl_include_dir, config.sycl_include + '/sycl/') ) ) llvm_config.add_tool_substitutions(['llvm-spirv'], [config.sycl_tools_dir]) diff --git a/sycl/test/separate-compile/test.cpp b/sycl/test/separate-compile/test.cpp index 30e58f50aee9b..df97ae177d5d1 100644 --- a/sycl/test/separate-compile/test.cpp +++ b/sycl/test/separate-compile/test.cpp @@ -7,13 +7,13 @@ // >> host compilation... // Driver automatically adds -D_DLL and -D_MT on Windows with -fsycl. // Add those defines required for Windows and harmless for Linux. -// RUN: %clangxx -D_DLL -D_MT -std=c++17 -include sycl_ihdr_a.h -g -c %s -o a.o -I %sycl_include -Wno-sycl-strict +// RUN: %clangxx -D_DLL -D_MT -std=c++17 -include sycl_ihdr_a.h -g -c %s -o a.o -I %sycl_include/sycl -Wno-sycl-strict // // >> ---- compile src2 // >> device compilation... // RUN: %clangxx -DB_CPP=1 -fsycl-device-only -Xclang -fsycl-int-header=sycl_ihdr_b.h %s -c -o b_kernel.bc -Wno-sycl-strict // >> host compilation... -// RUN: %clangxx -DB_CPP=1 -D_DLL -D_MT -std=c++17 -include sycl_ihdr_b.h -g -c %s -o b.o -I %sycl_include -Wno-sycl-strict +// RUN: %clangxx -DB_CPP=1 -D_DLL -D_MT -std=c++17 -include sycl_ihdr_b.h -g -c %s -o b.o -I %sycl_include/sycl -Wno-sycl-strict // // >> ---- bundle .o with .spv // >> run bundler diff --git a/sycl/test/warnings/sycl_2020_deprecations.cpp b/sycl/test/warnings/sycl_2020_deprecations.cpp index 023b4b9d9f905..2c68781642059 100644 --- a/sycl/test/warnings/sycl_2020_deprecations.cpp +++ b/sycl/test/warnings/sycl_2020_deprecations.cpp @@ -1,5 +1,7 @@ // RUN: %clangxx %fsycl-host-only -fsyntax-only -sycl-std=2020 -Xclang -verify -Xclang -verify-ignore-unexpected=note %s -o %t.out // RUN: %clangxx %fsycl-host-only -fsyntax-only -Xclang -verify -Xclang -verify-ignore-unexpected=note %s -o %t.out +// RUN: %clangxx %fsycl-host-only -fsyntax-only -sycl-std=2017 -Xclang -verify -Xclang -verify-ignore-unexpected=note %s -o %t.out +// RUN: %clangxx %fsycl-host-only -fsyntax-only -sycl-std=1.2.1 -Xclang -verify -Xclang -verify-ignore-unexpected=note %s -o %t.out #include From b69f889e7241d09db8909931c2277536f97ff019 Mon Sep 17 00:00:00 2001 From: Alexey Sachkov Date: Fri, 25 Jun 2021 16:22:52 +0300 Subject: [PATCH 15/21] Fix UB related to get_spec_constant_symbolic_ID Added a wrapper, which is defined in integration footer only after we define all specializations for `get_spec_constant_symbolic_ID`. By using it we ensure that we do not try to instantiate `get_spec_constant_symbolic_ID` until after we made all required specializations. --- clang/lib/Sema/SemaSYCL.cpp | 4 +++- clang/test/CodeGenSYCL/anonymous_integration_footer.cpp | 3 ++- clang/test/CodeGenSYCL/anonymous_integration_footer2.cpp | 3 ++- clang/test/CodeGenSYCL/integration_footer.cpp | 3 ++- sycl/include/CL/sycl/detail/kernel_desc.hpp | 7 +++++++ sycl/include/CL/sycl/detail/spec_const_integration.hpp | 8 ++------ sycl/include/CL/sycl/kernel_bundle.hpp | 9 ++++++--- sycl/unittests/spec_constants/DefaultValues.cpp | 2 +- 8 files changed, 25 insertions(+), 14 deletions(-) diff --git a/clang/lib/Sema/SemaSYCL.cpp b/clang/lib/Sema/SemaSYCL.cpp index 5fe6a8c48f8b5..8b952d9f773d5 100644 --- a/clang/lib/Sema/SemaSYCL.cpp +++ b/clang/lib/Sema/SemaSYCL.cpp @@ -4986,7 +4986,6 @@ bool SYCLIntegrationFooter::emit(raw_ostream &OS) { Policy.SuppressUnwrittenScope = true; OS << "#include \n"; - OS << "#include \n"; llvm::SmallSet VisitedSpecConstants; @@ -5029,6 +5028,9 @@ bool SYCLIntegrationFooter::emit(raw_ostream &OS) { OS << "} // namespace sycl\n"; OS << "} // __SYCL_INLINE_NAMESPACE(cl)\n"; } + + OS << "#include \n"; + return true; } diff --git a/clang/test/CodeGenSYCL/anonymous_integration_footer.cpp b/clang/test/CodeGenSYCL/anonymous_integration_footer.cpp index 8af6745e9bef1..54d42b8da80b2 100644 --- a/clang/test/CodeGenSYCL/anonymous_integration_footer.cpp +++ b/clang/test/CodeGenSYCL/anonymous_integration_footer.cpp @@ -10,7 +10,6 @@ int main() { } // CHECK: #include -// CHECK-NEXT: #include using namespace cl; @@ -402,3 +401,5 @@ constexpr sycl::specialization_id same_name{17}; // CHECK-NEXT: } // __SYCL_INLINE_NAMESPACE(cl) } } // namespace outer + +// CHECK: #include diff --git a/clang/test/CodeGenSYCL/anonymous_integration_footer2.cpp b/clang/test/CodeGenSYCL/anonymous_integration_footer2.cpp index 304f2a259b95c..0f627a15aa65e 100644 --- a/clang/test/CodeGenSYCL/anonymous_integration_footer2.cpp +++ b/clang/test/CodeGenSYCL/anonymous_integration_footer2.cpp @@ -10,7 +10,6 @@ int main() { } // CHECK: #include -// CHECK-NEXT: #include using namespace cl; @@ -155,3 +154,5 @@ constexpr sycl::specialization_id same_name{209}; // CHECK-NEXT: } // __SYCL_INLINE_NAMESPACE(cl) } } + +// CHECK: #include diff --git a/clang/test/CodeGenSYCL/integration_footer.cpp b/clang/test/CodeGenSYCL/integration_footer.cpp index 3818e94d94dbb..cbc26ac7f2706 100644 --- a/clang/test/CodeGenSYCL/integration_footer.cpp +++ b/clang/test/CodeGenSYCL/integration_footer.cpp @@ -8,7 +8,6 @@ int main() { } // CHECK: #include -// CHECK-NEXT: #include using namespace cl::sycl; @@ -190,3 +189,5 @@ auto x = HasVarTemplate::VarTempl.getDefaultValue(); // CHECK-NEXT: } // namespace detail // CHECK-NEXT: } // namespace sycl // CHECK-NEXT: } // __SYCL_INLINE_NAMESPACE(cl) + +// CHECK: #include diff --git a/sycl/include/CL/sycl/detail/kernel_desc.hpp b/sycl/include/CL/sycl/detail/kernel_desc.hpp index bf766e1d1deb7..ed1473648e5a5 100644 --- a/sycl/include/CL/sycl/detail/kernel_desc.hpp +++ b/sycl/include/CL/sycl/detail/kernel_desc.hpp @@ -50,10 +50,17 @@ template struct SpecConstantInfo { static constexpr const char *getName() { return ""; } }; +// This guard is needed because the libsycl.so can be compiled with C++ <=14 +// while the code requires C++17. This code is not supposed to be used by the +// libsycl.so so it should not be a problem. #if __cplusplus >= 201703L // Translates SYCL 2020 specialization constant type to its name. // Definition is in spec_const_integration.hpp template const char *get_spec_constant_symbolic_ID(); +// Wrapper is needed to delay instantiation of 'get_spec_constant_symbolic_ID' +// until after we have encountered all specializations for it generated by the +// compiler in integration footer. +template const char *get_spec_constant_symbolic_ID_wrapper(); #endif #ifndef __SYCL_UNNAMED_LAMBDA__ diff --git a/sycl/include/CL/sycl/detail/spec_const_integration.hpp b/sycl/include/CL/sycl/detail/spec_const_integration.hpp index 6025d35a5950d..9aea5fe8d3a66 100644 --- a/sycl/include/CL/sycl/detail/spec_const_integration.hpp +++ b/sycl/include/CL/sycl/detail/spec_const_integration.hpp @@ -17,12 +17,8 @@ namespace detail { #if __cplusplus >= 201703L // Translates SYCL 2020 specialization constant type to its name. -template const char *get_spec_constant_symbolic_ID() { -#ifdef SYCL_LANGUAGE_VERSION - return __builtin_sycl_unique_stable_id(SpecName); -#else - return ""; -#endif +template const char *get_spec_constant_symbolic_ID_wrapper() { + return get_spec_constant_symbolic_ID(); } #endif diff --git a/sycl/include/CL/sycl/kernel_bundle.hpp b/sycl/include/CL/sycl/kernel_bundle.hpp index 9eb256b63d2ab..151dc7d31d94a 100644 --- a/sycl/include/CL/sycl/kernel_bundle.hpp +++ b/sycl/include/CL/sycl/kernel_bundle.hpp @@ -263,7 +263,8 @@ class kernel_bundle : public detail::kernel_bundle_plain { /// \returns true if any device image in the kernel_bundle uses specialization /// constant whose address is SpecName template bool has_specialization_constant() const noexcept { - const char *SpecSymName = detail::get_spec_constant_symbolic_ID(); + const char *SpecSymName = + detail::get_spec_constant_symbolic_ID_wrapper(); return has_specialization_constant_impl(SpecSymName); } @@ -274,7 +275,8 @@ class kernel_bundle : public detail::kernel_bundle_plain { typename = detail::enable_if_t<_State == bundle_state::input>> void set_specialization_constant( typename std::remove_reference_t::value_type Value) { - const char *SpecSymName = detail::get_spec_constant_symbolic_ID(); + const char *SpecSymName = + detail::get_spec_constant_symbolic_ID_wrapper(); set_specialization_constant_impl(SpecSymName, &Value, sizeof(decltype(Value))); } @@ -284,7 +286,8 @@ class kernel_bundle : public detail::kernel_bundle_plain { template typename std::remove_reference_t::value_type get_specialization_constant() const { - const char *SpecSymName = detail::get_spec_constant_symbolic_ID(); + const char *SpecSymName = + detail::get_spec_constant_symbolic_ID_wrapper(); if (!is_specialization_constant_set(SpecSymName)) return SpecName.getDefaultValue(); diff --git a/sycl/unittests/spec_constants/DefaultValues.cpp b/sycl/unittests/spec_constants/DefaultValues.cpp index a5605124a533a..f330c12a12978 100644 --- a/sycl/unittests/spec_constants/DefaultValues.cpp +++ b/sycl/unittests/spec_constants/DefaultValues.cpp @@ -33,7 +33,7 @@ template <> struct KernelInfo { static constexpr bool callsAnyThisFreeFunction() { return false; } }; -template <> const char *get_spec_constant_symbolic_ID() { +template <> const char *get_spec_constant_symbolic_ID_wrapper() { return "SC1"; } } // namespace detail From 68c1f1ac6008bcf7fa625d18017773989c2eaccf Mon Sep 17 00:00:00 2001 From: Alexey Sachkov Date: Mon, 28 Jun 2021 10:10:07 +0300 Subject: [PATCH 16/21] Adjust sycl-offload-amdgcn.cpp test --- clang/test/Driver/sycl-offload-amdgcn.cpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/clang/test/Driver/sycl-offload-amdgcn.cpp b/clang/test/Driver/sycl-offload-amdgcn.cpp index 627932fc6f026..6824521c9fbe8 100644 --- a/clang/test/Driver/sycl-offload-amdgcn.cpp +++ b/clang/test/Driver/sycl-offload-amdgcn.cpp @@ -13,7 +13,7 @@ // CHK-ACTIONS: clang-offload-wrapper"{{.*}} "-host=x86_64-unknown-linux-gnu" "-target=amdgcn" "-kind=sycl"{{.*}} /// Check phases w/out specifying a compute capability. -// RUN: %clangxx -ccc-print-phases -std=c++11 -target x86_64-unknown-linux-gnu -fsycl -fsycl-use-footer \ +// RUN: %clangxx -ccc-print-phases -std=c++11 -target x86_64-unknown-linux-gnu -fsycl \ // RUN: -fsycl-targets=amdgcn-amd-amdhsa-sycldevice -mcpu=gfx906 %s 2>&1 \ // RUN: | FileCheck -check-prefix=CHK-PHASES-NO-CC %s // CHK-PHASES-NO-CC: 0: input, "{{.*}}", c++, (host-sycl) From 37be5406b3e7162d0e9a48d707db8794b2ae1701 Mon Sep 17 00:00:00 2001 From: Alexey Sachkov Date: Mon, 28 Jun 2021 10:40:44 +0300 Subject: [PATCH 17/21] Add a test for -fno-sycl-use-footer --- clang/test/Driver/sycl-int-footer.cpp | 6 ++++++ 1 file changed, 6 insertions(+) diff --git a/clang/test/Driver/sycl-int-footer.cpp b/clang/test/Driver/sycl-int-footer.cpp index 7a7fc10499c84..cdde05c044ca6 100644 --- a/clang/test/Driver/sycl-int-footer.cpp +++ b/clang/test/Driver/sycl-int-footer.cpp @@ -24,3 +24,9 @@ // FOOTER_PREPROC_USE: clang{{.*}} "-fsycl-is-device"{{.*}} "[[DEVICE_PP]]" // FOOTER_PREPROC_USE: clang-offload-bundler{{.*}} "-outputs=[[HOST_PP:.+\.ii]],[[DEVICE1:.+\.ii]]" // FOOTER_PREPROC_USE: clang{{.*}} "-fsycl-is-host"{{.*}} "[[HOST_PP]]" + +/// Check that integration footer can be disabled +// RUN: %clangxx -fsycl -fno-sycl-use-footer %s -### 2>&1 \ +// RUN: | FileCheck -check-prefix NO-FOOTER --implicit-check-not "-fsycl-int-footer" %s +// NO-FOOTER: clang{{.*}} "-fsycl-is-device"{{.*}} "-fsycl-int-header=[[INTHEADER:.+\.h]]" "-sycl-std={{.*}}" +// NO-FOOTER: clang{{.*}} "-include" "[[INTHEADER]]"{{.*}} "-fsycl-is-host"{{.*}} "-o" From 48e59c917ecb6795632b1e981bd9593f96a34505 Mon Sep 17 00:00:00 2001 From: Alexey Sachkov Date: Mon, 28 Jun 2021 11:02:58 +0300 Subject: [PATCH 18/21] Rename get_spec_constant_symbolic_ID The primary function/wrapper which is used by DPC++ RT is left as-is, i.e. `get_spec_constant_symbolic_ID`. The internal function and its specialziations which are generated by the compiler renamed to `get_spec_constant_symbolic_ID_impl`. --- clang/lib/Sema/SemaSYCL.cpp | 2 +- .../anonymous_integration_footer.cpp | 40 +++++++++---------- .../anonymous_integration_footer2.cpp | 18 ++++----- clang/test/CodeGenSYCL/integration_footer.cpp | 24 +++++------ sycl/include/CL/sycl/detail/kernel_desc.hpp | 14 ++++--- .../CL/sycl/detail/spec_const_integration.hpp | 4 +- sycl/include/CL/sycl/kernel_bundle.hpp | 6 +-- .../spec_constants/DefaultValues.cpp | 2 +- 8 files changed, 56 insertions(+), 54 deletions(-) diff --git a/clang/lib/Sema/SemaSYCL.cpp b/clang/lib/Sema/SemaSYCL.cpp index 8eca0e83394c1..43552b63de3fc 100644 --- a/clang/lib/Sema/SemaSYCL.cpp +++ b/clang/lib/Sema/SemaSYCL.cpp @@ -4990,7 +4990,7 @@ bool SYCLIntegrationFooter::emit(raw_ostream &OS) { OS << "namespace sycl {\n"; OS << "namespace detail {\n"; OS << "template<>\n"; - OS << "inline const char *get_spec_constant_symbolic_ID<"; + OS << "inline const char *get_spec_constant_symbolic_ID_impl<"; if (VD->isInAnonymousNamespace()) { OS << TopShim; diff --git a/clang/test/CodeGenSYCL/anonymous_integration_footer.cpp b/clang/test/CodeGenSYCL/anonymous_integration_footer.cpp index 54d42b8da80b2..ec9f726bd7b50 100644 --- a/clang/test/CodeGenSYCL/anonymous_integration_footer.cpp +++ b/clang/test/CodeGenSYCL/anonymous_integration_footer.cpp @@ -21,7 +21,7 @@ struct S1 { // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> - // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::S1::a>() { + // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::S1::a>() { // CHECK-NEXT: return "_ZN2S11aE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -33,7 +33,7 @@ constexpr sycl::specialization_id b{2}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::b>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::b>() { // CHECK-NEXT: return "____ZL1b"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -44,7 +44,7 @@ inline constexpr sycl::specialization_id c{3}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::c>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::c>() { // CHECK-NEXT: return "_Z1c"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -55,7 +55,7 @@ static constexpr sycl::specialization_id d{4}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::d>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::d>() { // CHECK-NEXT: return "____ZL1d"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -76,7 +76,7 @@ struct S2 { // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> - // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { + // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { // CHECK-NEXT: return "____ZN12_GLOBAL__N_12S21aE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -94,7 +94,7 @@ template class S3<1>; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::S3<1>::a>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::S3<1>::a>() { // CHECK-NEXT: return "_ZN2S3ILi1EE1aE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -105,7 +105,7 @@ template class S3<2>; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::S3<2>::a>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::S3<2>::a>() { // CHECK-NEXT: return "_ZN2S3ILi2EE1aE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -118,7 +118,7 @@ constexpr sycl::specialization_id same_name{5}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::inner::same_name>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::inner::same_name>() { // CHECK-NEXT: return "____ZN5innerL9same_nameE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -130,7 +130,7 @@ constexpr sycl::specialization_id same_name{6}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::same_name>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::same_name>() { // CHECK-NEXT: return "____ZL9same_name"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -149,7 +149,7 @@ constexpr sycl::specialization_id same_name{7}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { // CHECK-NEXT: return "____ZN12_GLOBAL__N_19same_nameE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -170,7 +170,7 @@ constexpr sycl::specialization_id same_name{8}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { // CHECK-NEXT: return "____ZN12_GLOBAL__N_15inner9same_nameE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -194,7 +194,7 @@ constexpr sycl::specialization_id same_name{9}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::inner::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::inner::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { // CHECK-NEXT: return "____ZN5inner12_GLOBAL__N_19same_nameE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -209,7 +209,7 @@ constexpr sycl::specialization_id same_name{10}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::outer::same_name>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::outer::same_name>() { // CHECK-NEXT: return "____ZN5outerL9same_nameE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -230,7 +230,7 @@ constexpr sycl::specialization_id same_name{11}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::outer::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::outer::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { // CHECK-NEXT: return "____ZN5outer12_GLOBAL__N_19same_nameE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -252,7 +252,7 @@ constexpr sycl::specialization_id same_name{12}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::outer::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::outer::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { // CHECK-NEXT: return "____ZN5outer12_GLOBAL__N_15inner9same_nameE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -290,7 +290,7 @@ constexpr sycl::specialization_id same_name{13}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::outer::__sycl_detail::__spec_id_shim_[[SHIM_ID_2]]()>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::outer::__sycl_detail::__spec_id_shim_[[SHIM_ID_2]]()>() { // CHECK-NEXT: return "____ZN5outer12_GLOBAL__N_15inner12_GLOBAL__N_19same_nameE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -315,7 +315,7 @@ constexpr sycl::specialization_id same_name{14}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { // CHECK-NEXT: return "____ZN12_GLOBAL__N_15outer9same_nameE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -345,7 +345,7 @@ constexpr sycl::specialization_id same_name{15}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::__sycl_detail::__spec_id_shim_[[SHIM_ID2]]()>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::__sycl_detail::__spec_id_shim_[[SHIM_ID2]]()>() { // CHECK-NEXT: return "____ZN12_GLOBAL__N_15outer12_GLOBAL__N_19same_nameE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -375,7 +375,7 @@ constexpr sycl::specialization_id same_name{16}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::__sycl_detail::__spec_id_shim_[[SHIM_ID2]]()>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::__sycl_detail::__spec_id_shim_[[SHIM_ID2]]()>() { // CHECK-NEXT: return "____ZN12_GLOBAL__N_15outer12_GLOBAL__N_15inner9same_nameE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -393,7 +393,7 @@ constexpr sycl::specialization_id same_name{17}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::outer::inner::same_name>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::outer::inner::same_name>() { // CHECK-NEXT: return "____ZN5outer5innerL9same_nameE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail diff --git a/clang/test/CodeGenSYCL/anonymous_integration_footer2.cpp b/clang/test/CodeGenSYCL/anonymous_integration_footer2.cpp index 0f627a15aa65e..d90f593c4043b 100644 --- a/clang/test/CodeGenSYCL/anonymous_integration_footer2.cpp +++ b/clang/test/CodeGenSYCL/anonymous_integration_footer2.cpp @@ -20,7 +20,7 @@ struct S1 { // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::S1::a>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::S1::a>() { // CHECK-NEXT: return "_ZN2S11aE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -32,7 +32,7 @@ constexpr sycl::specialization_id b{202}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::b>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::b>() { // CHECK-NEXT: return "____ZL1b"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -43,7 +43,7 @@ inline constexpr sycl::specialization_id c{3}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::c>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::c>() { // CHECK-NEXT: return "_Z1c"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -54,7 +54,7 @@ static constexpr sycl::specialization_id d{205}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::d>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::d>() { // CHECK-NEXT: return "____ZL1d"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -67,7 +67,7 @@ constexpr sycl::specialization_id same_name{5}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::inner::same_name>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::inner::same_name>() { // CHECK-NEXT: return "____ZN5innerL9same_nameE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -79,7 +79,7 @@ constexpr sycl::specialization_id same_name{6}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::same_name>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::same_name>() { // CHECK-NEXT: return "____ZL9same_name"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -98,7 +98,7 @@ constexpr sycl::specialization_id same_name{207}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { // CHECK-NEXT: return "____ZN12_GLOBAL__N_19same_nameE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -119,7 +119,7 @@ constexpr sycl::specialization_id same_name{208}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { // CHECK-NEXT: return "____ZN12_GLOBAL__N_15inner9same_nameE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -146,7 +146,7 @@ constexpr sycl::specialization_id same_name{209}; // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::outer::inner::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::outer::inner::__sycl_detail::__spec_id_shim_[[SHIM_ID]]()>() { // CHECK-NEXT: return "____ZN5outer5inner12_GLOBAL__N_19same_nameE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail diff --git a/clang/test/CodeGenSYCL/integration_footer.cpp b/clang/test/CodeGenSYCL/integration_footer.cpp index cbc26ac7f2706..5cacc62004f58 100644 --- a/clang/test/CodeGenSYCL/integration_footer.cpp +++ b/clang/test/CodeGenSYCL/integration_footer.cpp @@ -15,7 +15,7 @@ cl::sycl::specialization_id GlobalSpecID; // CHECK: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::GlobalSpecID>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::GlobalSpecID>() { // CHECK-NEXT: return "_Z12GlobalSpecID"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -26,7 +26,7 @@ struct Wrapper { // CHECK: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> - // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::Wrapper::WrapperSpecID>() { + // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::Wrapper::WrapperSpecID>() { // CHECK-NEXT: return "_ZN7Wrapper13WrapperSpecIDE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -41,7 +41,7 @@ template class WrapperTemplate; // CHECK: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::WrapperTemplate::WrapperSpecID>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::WrapperTemplate::WrapperSpecID>() { // CHECK-NEXT: return "_ZN15WrapperTemplateIiE13WrapperSpecIDE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -50,7 +50,7 @@ template class WrapperTemplate; // CHECK: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::WrapperTemplate::WrapperSpecID>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::WrapperTemplate::WrapperSpecID>() { // CHECK-NEXT: return "_ZN15WrapperTemplateIdE13WrapperSpecIDE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -61,7 +61,7 @@ specialization_id NSSpecID; // CHECK: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::Foo::NSSpecID>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::Foo::NSSpecID>() { // CHECK-NEXT: return "_ZN3Foo8NSSpecIDE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -71,7 +71,7 @@ specialization_id InlineNSSpecID; // CHECK: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::Foo::InlineNSSpecID>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::Foo::InlineNSSpecID>() { // CHECK-NEXT: return "_ZN3Foo3Bar14InlineNSSpecIDE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -80,7 +80,7 @@ specialization_id NSSpecID; // CHECK: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::Foo::Bar::NSSpecID>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::Foo::Bar::NSSpecID>() { // CHECK-NEXT: return "_ZN3Foo3Bar8NSSpecIDE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -91,7 +91,7 @@ struct Wrapper { // CHECK: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> - // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::Foo::Wrapper::WrapperSpecID>() { + // CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::Foo::Wrapper::WrapperSpecID>() { // CHECK-NEXT: return "_ZN3Foo3Bar7Wrapper13WrapperSpecIDE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -106,7 +106,7 @@ template class WrapperTemplate; // CHECK: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::Foo::WrapperTemplate::WrapperSpecID>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::Foo::WrapperTemplate::WrapperSpecID>() { // CHECK-NEXT: return "_ZN3Foo3Bar15WrapperTemplateIiE13WrapperSpecIDE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -115,7 +115,7 @@ template class WrapperTemplate; // CHECK: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::Foo::WrapperTemplate::WrapperSpecID>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::Foo::WrapperTemplate::WrapperSpecID>() { // CHECK-NEXT: return "_ZN3Foo3Bar15WrapperTemplateIdE13WrapperSpecIDE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -137,7 +137,7 @@ specialization_id AnonNSSpecID; // CHECK: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::Foo::__sycl_detail::__spec_id_shim_[[SHIM0]]()>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::Foo::__sycl_detail::__spec_id_shim_[[SHIM0]]()>() { // CHECK-NEXT: return "____ZN3Foo12_GLOBAL__N_112AnonNSSpecIDE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail @@ -183,7 +183,7 @@ auto x = HasVarTemplate::VarTempl.getDefaultValue(); // CHECK-NEXT: namespace sycl { // CHECK-NEXT: namespace detail { // CHECK-NEXT: template<> -// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID<::__sycl_detail::__spec_id_shim_[[SHIM1]]()>() { +// CHECK-NEXT: inline const char *get_spec_constant_symbolic_ID_impl<::__sycl_detail::__spec_id_shim_[[SHIM1]]()>() { // CHECK-NEXT: return "____ZN12_GLOBAL__N_114HasVarTemplate8VarTemplIiLi2EEE"; // CHECK-NEXT: } // CHECK-NEXT: } // namespace detail diff --git a/sycl/include/CL/sycl/detail/kernel_desc.hpp b/sycl/include/CL/sycl/detail/kernel_desc.hpp index ed1473648e5a5..3182e9c5de431 100644 --- a/sycl/include/CL/sycl/detail/kernel_desc.hpp +++ b/sycl/include/CL/sycl/detail/kernel_desc.hpp @@ -54,13 +54,15 @@ template struct SpecConstantInfo { // while the code requires C++17. This code is not supposed to be used by the // libsycl.so so it should not be a problem. #if __cplusplus >= 201703L -// Translates SYCL 2020 specialization constant type to its name. -// Definition is in spec_const_integration.hpp +// Translates SYCL 2020 `specialization_id` to a unique symbolic identifier. +// There are no primary definition, only specializations in the integration +// footer. +template const char *get_spec_constant_symbolic_ID_impl(); +// Wrapper is needed to delay instantiation of +// 'get_spec_constant_symbolic_ID_impl' until after we have encountered all +// specializations for it generated by the compiler in the integration footer. +// Definition in spec_const_integration.hpp. template const char *get_spec_constant_symbolic_ID(); -// Wrapper is needed to delay instantiation of 'get_spec_constant_symbolic_ID' -// until after we have encountered all specializations for it generated by the -// compiler in integration footer. -template const char *get_spec_constant_symbolic_ID_wrapper(); #endif #ifndef __SYCL_UNNAMED_LAMBDA__ diff --git a/sycl/include/CL/sycl/detail/spec_const_integration.hpp b/sycl/include/CL/sycl/detail/spec_const_integration.hpp index 9aea5fe8d3a66..05d666ebcd7b1 100644 --- a/sycl/include/CL/sycl/detail/spec_const_integration.hpp +++ b/sycl/include/CL/sycl/detail/spec_const_integration.hpp @@ -17,8 +17,8 @@ namespace detail { #if __cplusplus >= 201703L // Translates SYCL 2020 specialization constant type to its name. -template const char *get_spec_constant_symbolic_ID_wrapper() { - return get_spec_constant_symbolic_ID(); +template const char *get_spec_constant_symbolic_ID() { + return get_spec_constant_symbolic_ID_impl(); } #endif diff --git a/sycl/include/CL/sycl/kernel_bundle.hpp b/sycl/include/CL/sycl/kernel_bundle.hpp index 151dc7d31d94a..f18dc3a31d21c 100644 --- a/sycl/include/CL/sycl/kernel_bundle.hpp +++ b/sycl/include/CL/sycl/kernel_bundle.hpp @@ -264,7 +264,7 @@ class kernel_bundle : public detail::kernel_bundle_plain { /// constant whose address is SpecName template bool has_specialization_constant() const noexcept { const char *SpecSymName = - detail::get_spec_constant_symbolic_ID_wrapper(); + detail::get_spec_constant_symbolic_ID(); return has_specialization_constant_impl(SpecSymName); } @@ -276,7 +276,7 @@ class kernel_bundle : public detail::kernel_bundle_plain { void set_specialization_constant( typename std::remove_reference_t::value_type Value) { const char *SpecSymName = - detail::get_spec_constant_symbolic_ID_wrapper(); + detail::get_spec_constant_symbolic_ID(); set_specialization_constant_impl(SpecSymName, &Value, sizeof(decltype(Value))); } @@ -287,7 +287,7 @@ class kernel_bundle : public detail::kernel_bundle_plain { typename std::remove_reference_t::value_type get_specialization_constant() const { const char *SpecSymName = - detail::get_spec_constant_symbolic_ID_wrapper(); + detail::get_spec_constant_symbolic_ID(); if (!is_specialization_constant_set(SpecSymName)) return SpecName.getDefaultValue(); diff --git a/sycl/unittests/spec_constants/DefaultValues.cpp b/sycl/unittests/spec_constants/DefaultValues.cpp index f330c12a12978..a5605124a533a 100644 --- a/sycl/unittests/spec_constants/DefaultValues.cpp +++ b/sycl/unittests/spec_constants/DefaultValues.cpp @@ -33,7 +33,7 @@ template <> struct KernelInfo { static constexpr bool callsAnyThisFreeFunction() { return false; } }; -template <> const char *get_spec_constant_symbolic_ID_wrapper() { +template <> const char *get_spec_constant_symbolic_ID() { return "SC1"; } } // namespace detail From b73a756a876121bd6667b83197d8815bd8e78be4 Mon Sep 17 00:00:00 2001 From: Dmitry Vodopyanov Date: Mon, 28 Jun 2021 15:25:11 +0300 Subject: [PATCH 19/21] Fix comment and clang-format --- sycl/include/CL/sycl/detail/spec_const_integration.hpp | 3 ++- sycl/include/CL/sycl/kernel_bundle.hpp | 9 +++------ 2 files changed, 5 insertions(+), 7 deletions(-) diff --git a/sycl/include/CL/sycl/detail/spec_const_integration.hpp b/sycl/include/CL/sycl/detail/spec_const_integration.hpp index 05d666ebcd7b1..056e197eeac72 100644 --- a/sycl/include/CL/sycl/detail/spec_const_integration.hpp +++ b/sycl/include/CL/sycl/detail/spec_const_integration.hpp @@ -16,7 +16,8 @@ namespace sycl { namespace detail { #if __cplusplus >= 201703L -// Translates SYCL 2020 specialization constant type to its name. +// Translates SYCL 2020 `specialization_id` to a unique symbolic identifier +// which is used internally by the toolchain template const char *get_spec_constant_symbolic_ID() { return get_spec_constant_symbolic_ID_impl(); } diff --git a/sycl/include/CL/sycl/kernel_bundle.hpp b/sycl/include/CL/sycl/kernel_bundle.hpp index f18dc3a31d21c..9eb256b63d2ab 100644 --- a/sycl/include/CL/sycl/kernel_bundle.hpp +++ b/sycl/include/CL/sycl/kernel_bundle.hpp @@ -263,8 +263,7 @@ class kernel_bundle : public detail::kernel_bundle_plain { /// \returns true if any device image in the kernel_bundle uses specialization /// constant whose address is SpecName template bool has_specialization_constant() const noexcept { - const char *SpecSymName = - detail::get_spec_constant_symbolic_ID(); + const char *SpecSymName = detail::get_spec_constant_symbolic_ID(); return has_specialization_constant_impl(SpecSymName); } @@ -275,8 +274,7 @@ class kernel_bundle : public detail::kernel_bundle_plain { typename = detail::enable_if_t<_State == bundle_state::input>> void set_specialization_constant( typename std::remove_reference_t::value_type Value) { - const char *SpecSymName = - detail::get_spec_constant_symbolic_ID(); + const char *SpecSymName = detail::get_spec_constant_symbolic_ID(); set_specialization_constant_impl(SpecSymName, &Value, sizeof(decltype(Value))); } @@ -286,8 +284,7 @@ class kernel_bundle : public detail::kernel_bundle_plain { template typename std::remove_reference_t::value_type get_specialization_constant() const { - const char *SpecSymName = - detail::get_spec_constant_symbolic_ID(); + const char *SpecSymName = detail::get_spec_constant_symbolic_ID(); if (!is_specialization_constant_set(SpecSymName)) return SpecName.getDefaultValue(); From 8e58b192848ed693414a1e18d05b9ab23d8cb5db Mon Sep 17 00:00:00 2001 From: Alexey Sachkov Date: Tue, 29 Jun 2021 14:08:11 +0300 Subject: [PATCH 20/21] Add test for compiler phases w/ and w/o integration footer --- clang/test/Driver/sycl-int-footer.cpp | 33 +++++++++++++++++++++++++++ 1 file changed, 33 insertions(+) diff --git a/clang/test/Driver/sycl-int-footer.cpp b/clang/test/Driver/sycl-int-footer.cpp index cdde05c044ca6..a1c8cd5df926e 100644 --- a/clang/test/Driver/sycl-int-footer.cpp +++ b/clang/test/Driver/sycl-int-footer.cpp @@ -30,3 +30,36 @@ // RUN: | FileCheck -check-prefix NO-FOOTER --implicit-check-not "-fsycl-int-footer" %s // NO-FOOTER: clang{{.*}} "-fsycl-is-device"{{.*}} "-fsycl-int-header=[[INTHEADER:.+\.h]]" "-sycl-std={{.*}}" // NO-FOOTER: clang{{.*}} "-include" "[[INTHEADER]]"{{.*}} "-fsycl-is-host"{{.*}} "-o" + +/// Check phases without integration footer +// RUN: %clangxx -fsycl -fno-sycl-device-lib=all -fno-sycl-use-footer %s -ccc-print-phases 2>&1 \ +// RUN: | FileCheck -check-prefix NO-FOOTER-PHASES -check-prefix COMMON-PHASES %s +// NO-FOOTER-PHASES: 0: input, "{{.*}}", c++, (host-sycl) +// NO-FOOTER-PHASES: [[#HOST_PREPROC:]]: preprocessor, {0}, c++-cpp-output, (host-sycl) +// NO-FOOTER-PHASES: 2: input, "{{.*}}", c++, (device-sycl) +// NO-FOOTER-PHASES: 3: preprocessor, {2}, c++-cpp-output, (device-sycl) +// NO-FOOTER-PHASES: [[#DEVICE_IR:]]: compiler, {3}, ir, (device-sycl) + +/// Check phases with integration footer +// RUN: %clangxx -fsycl -fno-sycl-device-lib=all %s -ccc-print-phases 2>&1 \ +// RUN: | FileCheck -check-prefix FOOTER-PHASES -check-prefix COMMON-PHASES %s +// FOOTER-PHASES: 0: input, "{{.*}}", c++, (host-sycl) +// FOOTER-PHASES: 1: preprocessor, {0}, c++-cpp-output, (host-sycl) +// FOOTER-PHASES: 2: append-footer, {1}, c++, (host-sycl) +// FOOTER-PHASES: [[#HOST_PREPROC:]]: preprocessor, {2}, c++-cpp-output, (host-sycl) +// FOOTER-PHASES: 4: input, "{{.*}}", c++, (device-sycl) +// FOOTER-PHASES: 5: preprocessor, {4}, c++-cpp-output, (device-sycl) +// FOOTER-PHASES: [[#DEVICE_IR:]]: compiler, {5}, ir, (device-sycl) + +// COMMON-PHASES: [[#OFFLOAD:]]: offload, "host-sycl (x86_64-unknown-linux-gnu)" {[[#HOST_PREPROC]]}, "device-sycl (spir64-unknown-unknown-sycldevice)" {[[#DEVICE_IR]]}, c++-cpp-output +// COMMON-PHASES: [[#OFFLOAD+1]]: compiler, {[[#OFFLOAD]]}, ir, (host-sycl) +// COMMON-PHASES: [[#OFFLOAD+2]]: backend, {[[#OFFLOAD+1]]}, assembler, (host-sycl) +// COMMON-PHASES: [[#OFFLOAD+3]]: assembler, {[[#OFFLOAD+2]]}, object, (host-sycl) +// COMMON-PHASES: [[#OFFLOAD+4]]: linker, {[[#OFFLOAD+3]]}, image, (host-sycl) +// COMMON-PHASES: [[#OFFLOAD+5]]: linker, {[[#DEVICE_IR]]}, ir, (device-sycl) +// COMMON-PHASES: [[#OFFLOAD+6]]: sycl-post-link, {[[#OFFLOAD+5]]}, tempfiletable, (device-sycl) +// COMMON-PHASES: [[#OFFLOAD+7]]: file-table-tform, {[[#OFFLOAD+6]]}, tempfilelist, (device-sycl) +// COMMON-PHASES: [[#OFFLOAD+8]]: llvm-spirv, {[[#OFFLOAD+7]]}, tempfilelist, (device-sycl) +// COMMON-PHASES: [[#OFFLOAD+9]]: file-table-tform, {[[#OFFLOAD+6]], [[#OFFLOAD+8]]}, tempfiletable, (device-sycl) +// COMMON-PHASES: [[#OFFLOAD+10]]: clang-offload-wrapper, {[[#OFFLOAD+9]]}, object, (device-sycl) +// COMMON-PHASES: [[#OFFLOAD+11]]: offload, "host-sycl (x86_64-unknown-linux-gnu)" {[[#OFFLOAD+4]]}, "device-sycl (spir64-unknown-unknown-sycldevice)" {[[#OFFLOAD+10]]}, image From 94764032bb96338bbbf8b37771459e44f28ff60b Mon Sep 17 00:00:00 2001 From: Alexey Sachkov Date: Tue, 29 Jun 2021 16:50:19 +0300 Subject: [PATCH 21/21] An attempt to make sycl-int-footer.cpp OS-independent --- clang/test/Driver/sycl-int-footer.cpp | 4 ++-- 1 file changed, 2 insertions(+), 2 deletions(-) diff --git a/clang/test/Driver/sycl-int-footer.cpp b/clang/test/Driver/sycl-int-footer.cpp index a1c8cd5df926e..cc54bf70f28a0 100644 --- a/clang/test/Driver/sycl-int-footer.cpp +++ b/clang/test/Driver/sycl-int-footer.cpp @@ -51,7 +51,7 @@ // FOOTER-PHASES: 5: preprocessor, {4}, c++-cpp-output, (device-sycl) // FOOTER-PHASES: [[#DEVICE_IR:]]: compiler, {5}, ir, (device-sycl) -// COMMON-PHASES: [[#OFFLOAD:]]: offload, "host-sycl (x86_64-unknown-linux-gnu)" {[[#HOST_PREPROC]]}, "device-sycl (spir64-unknown-unknown-sycldevice)" {[[#DEVICE_IR]]}, c++-cpp-output +// COMMON-PHASES: [[#OFFLOAD:]]: offload, "host-sycl (x86_64-{{.*}})" {[[#HOST_PREPROC]]}, "device-sycl (spir64-unknown-unknown-sycldevice)" {[[#DEVICE_IR]]}, c++-cpp-output // COMMON-PHASES: [[#OFFLOAD+1]]: compiler, {[[#OFFLOAD]]}, ir, (host-sycl) // COMMON-PHASES: [[#OFFLOAD+2]]: backend, {[[#OFFLOAD+1]]}, assembler, (host-sycl) // COMMON-PHASES: [[#OFFLOAD+3]]: assembler, {[[#OFFLOAD+2]]}, object, (host-sycl) @@ -62,4 +62,4 @@ // COMMON-PHASES: [[#OFFLOAD+8]]: llvm-spirv, {[[#OFFLOAD+7]]}, tempfilelist, (device-sycl) // COMMON-PHASES: [[#OFFLOAD+9]]: file-table-tform, {[[#OFFLOAD+6]], [[#OFFLOAD+8]]}, tempfiletable, (device-sycl) // COMMON-PHASES: [[#OFFLOAD+10]]: clang-offload-wrapper, {[[#OFFLOAD+9]]}, object, (device-sycl) -// COMMON-PHASES: [[#OFFLOAD+11]]: offload, "host-sycl (x86_64-unknown-linux-gnu)" {[[#OFFLOAD+4]]}, "device-sycl (spir64-unknown-unknown-sycldevice)" {[[#OFFLOAD+10]]}, image +// COMMON-PHASES: [[#OFFLOAD+11]]: offload, "host-sycl (x86_64-{{.*}})" {[[#OFFLOAD+4]]}, "device-sycl (spir64-unknown-unknown-sycldevice)" {[[#OFFLOAD+10]]}, image