Address review comments

Author: Victor Lomuller

…cl/include/sycl/detail/sycl_builtins.hpp …/sycl/detail/sycl_local_mem_builtins.hppsycl/include/sycl/detail/sycl_builtins.hpp renamed to sycl/include/sycl/detail/sycl_local_mem_builtins.hpp sycl/include/sycl/detail/sycl_builtins.hpp renamed to sycl/include/sycl/detail/sycl_local_mem_builtins.hpp

@@ -1,4 +1,4 @@
-//==---------------- sycl_builtins.hpp --- SYCL access ---------------------==//
+//==----- sycl_local_mem_builtins.hpp --- SYCL local memory builtins ------==//
 //
 // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
 // See https://llvm.org/LICENSE.txt for license information.

sycl/include/sycl/ext/oneapi/group_local_memory.hpp

@@ -9,11 +9,11 @@
 
 #include <sycl/access/access.hpp>             // for address_space, decorated
 #include <sycl/detail/defines_elementary.hpp> // for __SYCL_ALWAYS_INLINE
-#include <sycl/detail/sycl_builtins.hpp>      // for __sycl_allocateLocalMemory
-#include <sycl/detail/type_traits.hpp>        // for is_group
-#include <sycl/exception.hpp>                 // for exception
-#include <sycl/ext/intel/usm_pointers.hpp>    // for multi_ptr
-#include <sycl/group.hpp>                     // for workGroupBarrier
+#include <sycl/detail/sycl_local_mem_builtins.hpp> // for __sycl_allocateLocalMemory
+#include <sycl/detail/type_traits.hpp>             // for is_group
+#include <sycl/exception.hpp>                      // for exception
+#include <sycl/ext/intel/usm_pointers.hpp>         // for multi_ptr
+#include <sycl/group.hpp>                          // for workGroupBarrier
 
 #include <type_traits> // for enable_if_t
 

sycl/include/sycl/ext/oneapi/properties/property.hpp

@@ -211,7 +211,7 @@ enum PropKind : uint32_t {
   Prefetch = 76,
   Deterministic = 77,
   InitializeToIdentity = 78,
-  WorkGroupStaticMem = 79,
+  WorkGroupScratchSize = 79,
   // PropKindSize must always be the last value.
   PropKindSize = 80,
 };

sycl/include/sycl/ext/oneapi/work_group_scratch_memory.hpp

@@ -7,9 +7,9 @@
 //===----------------------------------------------------------------------===//
 #pragma once
 
-#include <sycl/detail/defines_elementary.hpp> // for __SYCL_ALWAYS_INLINE
-#include <sycl/detail/sycl_builtins.hpp>      // for __sycl_allocateLocalMemory
-#include <sycl/exception.hpp>                 // for exception
+#include <sycl/detail/defines_elementary.hpp>      // for __SYCL_ALWAYS_INLINE
+#include <sycl/detail/sycl_local_mem_builtins.hpp> // for __sycl_allocateLocalMemory
+#include <sycl/exception.hpp>                      // for exception
 #include <sycl/ext/oneapi/properties/properties.hpp> // for properties
 
 namespace sycl {
@@ -32,7 +32,7 @@ inline void *get_work_group_scratch_memory() {
 struct work_group_scratch_size
     : ::sycl::ext::oneapi::experimental::detail::run_time_property_key<
           work_group_scratch_size,
-          ::sycl::ext::oneapi::experimental::detail::WorkGroupStaticMem> {
+          ::sycl::ext::oneapi::experimental::detail::PropKind::WorkGroupScratchSize> {
   // Runtime property part
   constexpr work_group_scratch_size(size_t bytes) : size(bytes) {}
 
@@ -41,12 +41,6 @@ struct work_group_scratch_size
 
 using work_group_scratch_size_key = work_group_scratch_size;
 
-template <> struct is_property_key<work_group_scratch_size> : std::true_type {};
-
-template <>
-struct is_property_value<work_group_scratch_size>
-    : is_property_key<work_group_scratch_size> {};
-
 namespace detail {
 template <> struct PropertyMetaInfo<work_group_scratch_size> {
   static constexpr const char *name = "sycl-work-group-static";

sycl/include/sycl/ext/oneapi/work_group_static.hpp

@@ -34,7 +34,7 @@ template <typename T> class __SYCL_WG_SCOPE work_group_static final {
       std::is_trivially_destructible_v<T> &&
           std::is_trivially_constructible_v<T>,
       "Can only be used with trivially constructible and destructible types");
-  static_assert(!std::is_const_v<T> || !std::is_volatile_v<T>,
+  static_assert(!std::is_const_v<T> && !std::is_volatile_v<T>,
                 "Can only be used with non const and non volatile types");
   __SYCL_ALWAYS_INLINE work_group_static() = default;
   work_group_static(const work_group_static &) = delete;

sycl/test/include_deps/sycl_detail_core.hpp.cpp

@@ -159,6 +159,7 @@
 // CHECK-NEXT: ext/oneapi/experimental/virtual_functions.hpp
 // CHECK-NEXT: ext/oneapi/kernel_properties/properties.hpp
 // CHECK-NEXT: ext/oneapi/work_group_scratch_memory.hpp
+// CHECK-NEXT: detail/sycl_local_mem_builtins.hpp
 // CHECK-NEXT: kernel.hpp
 // CHECK-NEXT: kernel_bundle.hpp
 // CHECK-NEXT: ext/oneapi/experimental/free_function_traits.hpp