[SYCL][NFC] Refactor __SYCL_INLINE macro

The problem with previous definition of macro is that clang-format
insert additional indentation:

Original code:
```c++
__SYCL_INLINE namespace cl {
// some code
```

When applied clang-format:
```c++
__SYCL_INLINE namespace cl {
  // some code
```

Replaced `__SYCL_INLINE` with `__SYCL_INLINE_NAMESPACE(X)` which defines
either inline or regular namespace and marked it as `NamespaceMacro` in
clang-format config file to avoid dealing with broken formatting while
adding new code

Signed-off-by: Alexey Sachkov <alexey.sachkov@intel.com>

Author: AlexeySachkov

clang/lib/Sema/SemaSYCL.cpp

@@ -1672,7 +1672,7 @@ void SYCLIntegrationHeader::emit(raw_ostream &O) {
   }
   O << "\n";
 
-  O << "__SYCL_INLINE namespace cl {\n";
+  O << "__SYCL_INLINE_NAMESPACE(cl) {\n";
   O << "namespace sycl {\n";
   O << "namespace detail {\n";
 
@@ -1765,7 +1765,7 @@ void SYCLIntegrationHeader::emit(raw_ostream &O) {
   O << "\n";
   O << "} // namespace detail\n";
   O << "} // namespace sycl\n";
-  O << "} // namespace cl\n";
+  O << "} // __SYCL_INLINE_NAMESPACE(cl)\n";
   O << "\n";
 }
 

clang/test/CodeGenSYCL/wrapped-accessor.cpp

@@ -5,7 +5,7 @@
 
 // CHECK: class wrapped_access;
 
-// CHECK: namespace cl {
+// CHECK: __SYCL_INLINE_NAMESPACE(cl) {
 // CHECK-NEXT: namespace sycl {
 // CHECK-NEXT: namespace detail {
 

sycl/.clang-format

@@ -1,2 +1,3 @@
 BasedOnStyle: LLVM
 TypenameMacros: ['PI_CALL' ,'PI_CALL_THROW', 'PI_CALL_NOCHECK']
+NamespaceMacros: ['__SYCL_INLINE_NAMESPACE']

sycl/include/CL/sycl/access/access.hpp

@@ -9,7 +9,7 @@
 
 #include <CL/sycl/detail/defines.hpp>
 
-__SYCL_INLINE namespace cl {
+__SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
 namespace access {
 
@@ -180,4 +180,4 @@ struct remove_AS<__OPENCL_CONSTANT_AS__ T> {
 } // namespace detail
 
 }  // namespace sycl
-}  // namespace cl
+}  // __SYCL_INLINE_NAMESPACE(cl)

sycl/include/CL/sycl/accessor.hpp

@@ -143,7 +143,7 @@
 // accessor_common contains several helpers common for both accessor(1) and
 // accessor(3)
 
-__SYCL_INLINE namespace cl {
+__SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
 
 template <typename DataT, int Dimensions, access::mode AccessMode,
@@ -1264,7 +1264,7 @@ class accessor<DataT, Dimensions, AccessMode, access::target::image_array,
 };
 
 } // namespace sycl
-} // namespace cl
+} // __SYCL_INLINE_NAMESPACE(cl)
 
 namespace std {
 template <typename DataT, int Dimensions, cl::sycl::access::mode AccessMode,

sycl/include/CL/sycl/aliases.hpp

@@ -13,7 +13,7 @@
 #include <cstddef>
 #include <cstdint>
 
-__SYCL_INLINE namespace cl {
+__SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
 template <typename T, int N> class vec;
 namespace detail {
@@ -22,7 +22,7 @@ class half;
 } // namespace half_impl
 } // namespace detail
 } // namespace sycl
-} // namespace cl
+} // __SYCL_INLINE_NAMESPACE(cl)
 
 // FIXME: line below exports 'half' into global namespace, which seems incorrect
 // However, SYCL 1.2.1 spec considers 'half' to be a fundamental C++ data type
@@ -69,7 +69,7 @@ using half = cl::sycl::detail::half_impl::half;
   MAKE_VECTOR_ALIASES_FOR_OPENCL_TYPES(N)                                      \
   MAKE_VECTOR_ALIASES_FOR_SIGNED_AND_UNSIGNED_TYPES(N)
 
-__SYCL_INLINE namespace cl {
+__SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
 using byte = std::uint8_t;
 using schar = signed char;
@@ -101,7 +101,7 @@ MAKE_VECTOR_ALIASES_FOR_VECTOR_LENGTH(4)
 MAKE_VECTOR_ALIASES_FOR_VECTOR_LENGTH(8)
 MAKE_VECTOR_ALIASES_FOR_VECTOR_LENGTH(16)
 } // namespace sycl
-} // namespace cl
+} // __SYCL_INLINE_NAMESPACE(cl)
 
 #undef MAKE_VECTOR_ALIAS
 #undef MAKE_VECTOR_ALIASES_FOR_ARITHMETIC_TYPES

sycl/include/CL/sycl/atomic.hpp

@@ -23,7 +23,7 @@
   static_assert(!std::is_same<T, float>::value,                                \
                 "SYCL atomic function not available for float type")
 
-__SYCL_INLINE namespace cl {
+__SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
 
 enum class memory_order : int { relaxed };
@@ -62,11 +62,11 @@ template <> struct GetSpirvMemoryScope<access::address_space::local_space> {
 
 } // namespace detail
 } // namespace sycl
-} // namespace cl
+} // __SYCL_INLINE_NAMESPACE(cl)
 
 #ifndef __SYCL_DEVICE_ONLY__
 // host implementation of SYCL atomics
-__SYCL_INLINE namespace cl {
+__SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
 namespace detail {
 // Translate cl::sycl::memory_order or __spv::MemorySemanticsMask
@@ -81,7 +81,7 @@ static inline std::memory_order getStdMemoryOrder(::cl::sycl::memory_order MS) {
 }
 } // namespace detail
 } // namespace sycl
-} // namespace cl
+} // __SYCL_INLINE_NAMESPACE(cl)
 
 // std::atomic version of atomic SPIR-V builtins
 
@@ -161,7 +161,7 @@ extern T __spirv_AtomicMax(std::atomic<T> *Ptr, __spv::Scope S,
 
 #endif // !defined(__SYCL_DEVICE_ONLY__)
 
-__SYCL_INLINE namespace cl {
+__SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
 
 template <typename T, access::address_space addressSpace =
@@ -369,6 +369,6 @@ T atomic_fetch_max(atomic<T, addressSpace> Object, T Operand,
 }
 
 } // namespace sycl
-} // namespace cl
+} // __SYCL_INLINE_NAMESPACE(cl)
 
 #undef STATIC_ASSERT_NOT_FLOAT

sycl/include/CL/sycl/buffer.hpp

@@ -14,7 +14,7 @@
 
 // TODO: 4.3.4 Properties
 
-__SYCL_INLINE namespace cl {
+__SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
 class handler;
 class queue;
@@ -396,7 +396,7 @@ buffer(const T *, const range<dimensions> &, const property_list & = {})
 #endif // __cpp_deduction_guides
 
 } // namespace sycl
-} // namespace cl
+} // __SYCL_INLINE_NAMESPACE(cl)
 
 namespace std {
 template <typename T, int dimensions, typename AllocatorT>

sycl/include/CL/sycl/builtins.hpp

@@ -17,17 +17,17 @@
 // TODO Decide whether to mark functions with this attribute.
 #define __NOEXC /*noexcept*/
 
-__SYCL_INLINE namespace cl {
+__SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
 #ifdef __SYCL_DEVICE_ONLY__
 #define __sycl_std
 #else
 namespace __sycl_std = __host_std;
 #endif
 } // namespace sycl
-} // namespace cl
+} // __SYCL_INLINE_NAMESPACE(cl)
 
-__SYCL_INLINE namespace cl {
+__SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
 /* ----------------- 4.13.3 Math functions. ---------------------------------*/
 // genfloat acos (genfloat x)
@@ -1546,6 +1546,6 @@ detail::enable_if_t<detail::is_genfloatf<T>::value, T> tan(T x) __NOEXC {
 
 } // namespace half_precision
 } // namespace sycl
-} // namespace cl
+} // __SYCL_INLINE_NAMESPACE(cl)
 
 #undef __NOEXC

sycl/include/CL/sycl/context.hpp

@@ -14,7 +14,7 @@
 #include <type_traits>
 // 4.6.2 Context class
 
-__SYCL_INLINE namespace cl {
+__SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
 // Forward declarations
 class device;
@@ -139,7 +139,7 @@ class context {
 };
 
 } // namespace sycl
-} // namespace cl
+} // __SYCL_INLINE_NAMESPACE(cl)
 
 namespace std {
 template <> struct hash<cl::sycl::context> {