use detail namespace

Signed-off-by: Sarnie, Nick <[email protected]>

Author: sarnex

sycl/include/sycl/ext/intel/experimental/kernel_properties.hpp renamed to sycl/include/sycl/detail/kernel_properties.hpp

@@ -15,49 +15,43 @@
 
 namespace sycl {
 __SYCL_INLINE_VER_NAMESPACE(_V1) {
-namespace ext {
-namespace intel::experimental {
-
+namespace detail {
 enum class register_alloc_mode_enum : uint32_t {
   automatic = 0,
   large = 2,
 };
 
 struct register_alloc_mode_key {
   template <register_alloc_mode_enum Mode>
-  using value_t = oneapi::experimental::property_value<
+  using value_t = sycl::ext::oneapi::experimental::property_value<
       register_alloc_mode_key,
       std::integral_constant<register_alloc_mode_enum, Mode>>;
 };
 
 template <register_alloc_mode_enum Mode>
 inline constexpr register_alloc_mode_key::value_t<Mode> register_alloc_mode;
-} // namespace intel::experimental
+} // namespace detail
 
-namespace oneapi::experimental {
+namespace ext::oneapi::experimental {
 template <>
-struct is_property_key<intel::experimental::register_alloc_mode_key>
-    : std::true_type {};
+struct is_property_key<sycl::detail::register_alloc_mode_key> : std::true_type {
+};
 
 namespace detail {
-template <>
-struct PropertyToKind<intel::experimental::register_alloc_mode_key> {
+template <> struct PropertyToKind<sycl::detail::register_alloc_mode_key> {
   static constexpr PropKind Kind = PropKind::RegisterAllocMode;
 };
 
 template <>
-struct IsCompileTimeProperty<intel::experimental::register_alloc_mode_key>
+struct IsCompileTimeProperty<sycl::detail::register_alloc_mode_key>
     : std::true_type {};
 
-template <intel::experimental::register_alloc_mode_enum Mode>
-struct PropertyMetaInfo<
-    intel::experimental::register_alloc_mode_key::value_t<Mode>> {
+template <sycl::detail::register_alloc_mode_enum Mode>
+struct PropertyMetaInfo<sycl::detail::register_alloc_mode_key::value_t<Mode>> {
   static constexpr const char *name = "sycl-register-alloc-mode";
-  static constexpr intel::experimental::register_alloc_mode_enum value = Mode;
+  static constexpr sycl::detail::register_alloc_mode_enum value = Mode;
 };
-
 } // namespace detail
-} // namespace oneapi::experimental
-} // namespace ext
+} // namespace ext::oneapi::experimental
 } // __SYCL_INLINE_VER_NAMESPACE(_V1)
 } // namespace sycl

sycl/include/sycl/sycl.hpp

@@ -63,7 +63,6 @@
 #include <sycl/ext/oneapi/backend/level_zero.hpp>
 #endif
 #include <sycl/ext/codeplay/experimental/fusion_wrapper.hpp>
-#include <sycl/ext/intel/experimental/kernel_properties.hpp>
 #include <sycl/ext/intel/experimental/pipe_properties.hpp>
 #include <sycl/ext/intel/experimental/pipes.hpp>
 #include <sycl/ext/intel/usm_pointers.hpp>

sycl/source/detail/program_manager/program_manager.cpp

@@ -22,12 +22,12 @@
 #include <sycl/backend_types.hpp>
 #include <sycl/context.hpp>
 #include <sycl/detail/common.hpp>
+#include <sycl/detail/kernel_properties.hpp>
 #include <sycl/detail/os_util.hpp>
 #include <sycl/detail/type_traits.hpp>
 #include <sycl/detail/util.hpp>
 #include <sycl/device.hpp>
 #include <sycl/exception.hpp>
-#include <sycl/ext/intel/experimental/kernel_properties.hpp>
 #include <sycl/ext/oneapi/experimental/spec_constant.hpp>
 #include <sycl/stl.hpp>
 
@@ -382,18 +382,14 @@ static void appendCompileOptionsForRegAllocMode(std::string &CompileOpts,
   if (!Prop)
     return;
   uint32_t PropVal = DeviceBinaryProperty(Prop).asUint32();
-  if (PropVal ==
-      static_cast<uint32_t>(
-          ext::intel::experimental::register_alloc_mode_enum::large)) {
+  if (PropVal == static_cast<uint32_t>(register_alloc_mode_enum::large)) {
     if (!CompileOpts.empty())
       CompileOpts += " ";
     // This option works for both LO AND OCL backends.
     CompileOpts += IsEsimdImage ? "-doubleGRF" : "-ze-opt-large-register-file";
   }
   // TODO: Support Auto GRF for ESIMD once vc supports it.
-  if (PropVal ==
-          static_cast<uint32_t>(
-              ext::intel::experimental::register_alloc_mode_enum::automatic) &&
+  if (PropVal == static_cast<uint32_t>(register_alloc_mode_enum::automatic) &&
       !IsEsimdImage) {
     if (!CompileOpts.empty())
       CompileOpts += " ";

sycl/test-e2e/DeviceCodeSplit/grf.cpp

@@ -26,9 +26,11 @@
 // RUN: env SYCL_PROGRAM_COMPILE_OPTIONS="-g" SYCL_PI_TRACE=-1 %{run} %t.out 2>&1 | FileCheck %s --check-prefixes=CHECK,CHECK-AUTO-WITH-VAR
 #include "../helpers.hpp"
 #include <iostream>
+#include <sycl/detail/kernel_properties.hpp>
 #include <sycl/sycl.hpp>
 
 using namespace sycl;
+using namespace sycl::detail;
 using namespace sycl::ext::intel::experimental;
 using namespace sycl::ext::oneapi::experimental;
 

sycl/test-e2e/ESIMD/large-grf.cpp

@@ -24,10 +24,12 @@
 #include "esimd_test_utils.hpp"
 
 #include <iostream>
+#include <sycl/detail/kernel_properties.hpp>
 #include <sycl/ext/intel/esimd.hpp>
 #include <sycl/sycl.hpp>
 
 using namespace sycl;
+using namespace sycl::detail;
 using namespace sycl::ext::intel::esimd;
 using namespace sycl::ext::intel::experimental;
 using namespace sycl::ext::intel::experimental::esimd;
@@ -121,8 +123,7 @@ int main(void) {
     buffer<float, 1> bufa(A.data(), range<1>(Size));
     queue q(esimd_test::ESIMDSelector, esimd_test::createExceptionHandler());
     sycl::ext::oneapi::experimental::properties prop{
-        sycl::ext::intel::experimental::register_alloc_mode<
-            sycl::ext::intel::experimental::register_alloc_mode_enum::large>};
+        register_alloc_mode<register_alloc_mode_enum::large>};
     auto dev = q.get_device();
     std::cout << "Running on " << dev.get_info<info::device::name>() << "\n";