diff --git a/SYCL/Basic/built-ins/printf.cpp b/SYCL/Basic/built-ins/printf.cpp index 3e2e4a83de..528b0852ed 100644 --- a/SYCL/Basic/built-ins/printf.cpp +++ b/SYCL/Basic/built-ins/printf.cpp @@ -11,6 +11,14 @@ #include #include +/* https://github.com/intel/llvm/pull/2246: namespace compatibility mode*/ +__SYCL_INLINE_NAMESPACE(cl) { +namespace sycl { +namespace intel {}; +namespace ONEAPI { + using namespace cl::sycl::intel; +}}}; + using namespace cl::sycl; // According to OpenCL C spec, the format string must be in constant address @@ -40,7 +48,7 @@ int main() { Queue.submit([&](handler &CGH) { CGH.single_task([=]() { // String - intel::experimental::printf(format_hello_world); + ONEAPI::experimental::printf(format_hello_world); // Due to a bug in Intel CPU Runtime for OpenCL on Windows, information // printed using such format strings (without %-specifiers) might // appear in different order if output is redirected to a file or @@ -49,8 +57,8 @@ int main() { // CHECK: {{(Hello, World!)?}} // Integral types - intel::experimental::printf(format_int, (int32_t)123); - intel::experimental::printf(format_int, (int32_t)-123); + ONEAPI::experimental::printf(format_int, (int32_t)123); + ONEAPI::experimental::printf(format_int, (int32_t)-123); // CHECK: 123 // CHECK-NEXT: -123 @@ -59,8 +67,8 @@ int main() { // You can declare format string in non-global scope, but in this case // static keyword is required static const CONSTANT char format[] = "%f\n"; - intel::experimental::printf(format, 33.4f); - intel::experimental::printf(format, -33.4f); + ONEAPI::experimental::printf(format, 33.4f); + ONEAPI::experimental::printf(format, -33.4f); } // CHECK-NEXT: 33.4 // CHECK-NEXT: -33.4 @@ -72,21 +80,21 @@ int main() { using ocl_int4 = cl::sycl::vec::vector_t; { static const CONSTANT char format[] = "%v4d\n"; - intel::experimental::printf(format, (ocl_int4)v4); + ONEAPI::experimental::printf(format, (ocl_int4)v4); } // However, you are still able to print them by-element: { - intel::experimental::printf(format_vec, (int32_t)v4.w(), + ONEAPI::experimental::printf(format_vec, (int32_t)v4.w(), (int32_t)v4.z(), (int32_t)v4.y(), (int32_t)v4.x()); } #else // On host side you always have to print them by-element: - intel::experimental::printf(format_vec, (int32_t)v4.x(), + ONEAPI::experimental::printf(format_vec, (int32_t)v4.x(), (int32_t)v4.y(), (int32_t)v4.z(), (int32_t)v4.w()); - intel::experimental::printf(format_vec, (int32_t)v4.w(), + ONEAPI::experimental::printf(format_vec, (int32_t)v4.w(), (int32_t)v4.z(), (int32_t)v4.y(), (int32_t)v4.x()); #endif // __SYCL_DEVICE_ONLY__ @@ -99,7 +107,7 @@ int main() { // According to OpenCL spec, argument should be a void pointer { static const CONSTANT char format[] = "%p\n"; - intel::experimental::printf(format, (void *)Ptr); + ONEAPI::experimental::printf(format, (void *)Ptr); } // CHECK-NEXT: {{(0x)?[0-9a-fA-F]+$}} }); @@ -110,7 +118,7 @@ int main() { Queue.submit([&](handler &CGH) { CGH.parallel_for(range<1>(10), [=](id<1> i) { // cast to uint64_t to be sure that we pass 64-bit unsigned value - intel::experimental::printf(format_hello_world_2, (uint64_t)i.get(0)); + ONEAPI::experimental::printf(format_hello_world_2, (uint64_t)i.get(0)); }); }); Queue.wait(); diff --git a/SYCL/Basic/group-algorithm/all_of.cpp b/SYCL/Basic/group-algorithm/all_of.cpp index be37442d32..61e32e1397 100644 --- a/SYCL/Basic/group-algorithm/all_of.cpp +++ b/SYCL/Basic/group-algorithm/all_of.cpp @@ -11,8 +11,9 @@ #include #include #include +#include "ns_compat.h" using namespace sycl; -using namespace sycl::intel; +using namespace sycl::ONEAPI; template class all_of_kernel; diff --git a/SYCL/Basic/group-algorithm/any_of.cpp b/SYCL/Basic/group-algorithm/any_of.cpp index c9607e9159..fe4442402d 100644 --- a/SYCL/Basic/group-algorithm/any_of.cpp +++ b/SYCL/Basic/group-algorithm/any_of.cpp @@ -11,8 +11,9 @@ #include #include #include +#include "ns_compat.h" using namespace sycl; -using namespace sycl::intel; +using namespace sycl::ONEAPI; template class any_of_kernel; diff --git a/SYCL/Basic/group-algorithm/broadcast.cpp b/SYCL/Basic/group-algorithm/broadcast.cpp index 387ae8430c..a0b933c250 100644 --- a/SYCL/Basic/group-algorithm/broadcast.cpp +++ b/SYCL/Basic/group-algorithm/broadcast.cpp @@ -11,8 +11,9 @@ #include #include #include +#include "ns_compat.h" using namespace sycl; -using namespace sycl::intel; +using namespace sycl::ONEAPI; class broadcast_kernel; diff --git a/SYCL/Basic/group-algorithm/exclusive_scan.cpp b/SYCL/Basic/group-algorithm/exclusive_scan.cpp index 22d0644355..19b98585cf 100644 --- a/SYCL/Basic/group-algorithm/exclusive_scan.cpp +++ b/SYCL/Basic/group-algorithm/exclusive_scan.cpp @@ -13,8 +13,9 @@ #include #include #include +#include "ns_compat.h" using namespace sycl; -using namespace sycl::intel; +using namespace sycl::ONEAPI; template class exclusive_scan_kernel; diff --git a/SYCL/Basic/group-algorithm/inclusive_scan.cpp b/SYCL/Basic/group-algorithm/inclusive_scan.cpp index edea0142ef..8a99fa74d2 100644 --- a/SYCL/Basic/group-algorithm/inclusive_scan.cpp +++ b/SYCL/Basic/group-algorithm/inclusive_scan.cpp @@ -13,8 +13,9 @@ #include #include #include +#include "ns_compat.h" using namespace sycl; -using namespace sycl::intel; +using namespace sycl::ONEAPI; template class inclusive_scan_kernel; diff --git a/SYCL/Basic/group-algorithm/leader.cpp b/SYCL/Basic/group-algorithm/leader.cpp index f6c645f610..8e95763c38 100644 --- a/SYCL/Basic/group-algorithm/leader.cpp +++ b/SYCL/Basic/group-algorithm/leader.cpp @@ -9,8 +9,9 @@ #include #include +#include "ns_compat.h" using namespace sycl; -using namespace sycl::intel; +using namespace sycl::ONEAPI; class leader_kernel; diff --git a/SYCL/Basic/group-algorithm/none_of.cpp b/SYCL/Basic/group-algorithm/none_of.cpp index 51a68ab9c7..520492d0a3 100644 --- a/SYCL/Basic/group-algorithm/none_of.cpp +++ b/SYCL/Basic/group-algorithm/none_of.cpp @@ -11,8 +11,9 @@ #include #include #include +#include "ns_compat.h" using namespace sycl; -using namespace sycl::intel; +using namespace sycl::ONEAPI; template class none_of_kernel; diff --git a/SYCL/Basic/group-algorithm/ns_compat.h b/SYCL/Basic/group-algorithm/ns_compat.h new file mode 100644 index 0000000000..f4461031fe --- /dev/null +++ b/SYCL/Basic/group-algorithm/ns_compat.h @@ -0,0 +1,7 @@ +/* https://github.com/intel/llvm/pull/2231: namespace compatibility mode*/ +__SYCL_INLINE_NAMESPACE(cl) { +namespace sycl { +namespace intel {}; +namespace ONEAPI { + using namespace cl::sycl::intel; +}}}; diff --git a/SYCL/Basic/group-algorithm/reduce.cpp b/SYCL/Basic/group-algorithm/reduce.cpp index 10a458b019..1914159a93 100644 --- a/SYCL/Basic/group-algorithm/reduce.cpp +++ b/SYCL/Basic/group-algorithm/reduce.cpp @@ -12,8 +12,9 @@ #include #include #include +#include "ns_compat.h" using namespace sycl; -using namespace sycl::intel; +using namespace sycl::ONEAPI; template class reduce_kernel;