[SYCL] Implement diagnostic for !access_mode::read + const T (intel#8169)

mmoadeli · web-flow · commit aa69e4d9b86c · 2023-02-03T17:45:54.000-08:00
According to [Read only accessors](https://registry.khronos.org/SYCL/specs/sycl-2020/html/sycl-2020.html#_read_only_accessors) A const qualified DataT is only allowed for a read-only accessor. This PR: - Implements Implement diagnostic for !access_mode::read + const T - Adds test to for above diagnostic - Updates affected tests. - Assigns access::mode:read to access mode of local accessor when the data type is constant. The previous value was access::mode:read, which lead to an assert after implementing above diagnostic.
diff --git a/sycl/include/sycl/accessor.hpp b/sycl/include/sycl/accessor.hpp
@@ -314,6 +314,7 @@ class accessor_common {
       AccessMode == access::mode::discard_read_write;
 
   constexpr static bool IsAccessReadOnly = AccessMode == access::mode::read;
+  static constexpr bool IsConst = std::is_const<DataT>::value;
 
   constexpr static bool IsAccessReadWrite =
       AccessMode == access::mode::read_write;
@@ -371,6 +372,13 @@ class accessor_common {
   };
 };
 
+template <typename DataT> constexpr access::mode accessModeFromConstness() {
+  if constexpr (std::is_const<DataT>::value)
+    return access::mode::read;
+  else
+    return access::mode::read_write;
+}
+
 template <typename MayBeTag1, typename MayBeTag2>
 constexpr access::mode deduceAccessMode() {
   // property_list = {} is not properly detected by deduction guide,
@@ -1020,10 +1028,15 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(accessor) accessor :
   static constexpr bool IsGlobalBuf = AccessorCommonT::IsGlobalBuf;
   static constexpr bool IsHostBuf = AccessorCommonT::IsHostBuf;
   static constexpr bool IsPlaceH = AccessorCommonT::IsPlaceH;
+  static constexpr bool IsConst = AccessorCommonT::IsConst;
   template <int Dims>
   using AccessorSubscript =
       typename AccessorCommonT::template AccessorSubscript<Dims>;
 
+  static_assert(
+      !IsConst || IsAccessReadOnly,
+      "A const qualified DataT is only allowed for a read-only accessor");
+
   using ConcreteASPtrType = typename detail::DecoratedType<DataT, AS>::type *;
 
   using RefType = detail::const_if_const_AS<AS, DataT> &;
@@ -2392,7 +2405,13 @@ class __SYCL_SPECIAL_CLASS local_accessor_base :
                               access::target::local, IsPlaceholder>;
 
   using AccessorCommonT::AS;
-  using AccessorCommonT::IsAccessAnyWrite;
+
+  // Cannot do "using AccessorCommonT::Flag" as it doesn't work with g++ as host
+  // compiler, for some reason.
+  static constexpr bool IsAccessAnyWrite = AccessorCommonT::IsAccessAnyWrite;
+  static constexpr bool IsAccessReadOnly = AccessorCommonT::IsAccessReadOnly;
+  static constexpr bool IsConst = AccessorCommonT::IsConst;
+
   template <int Dims>
   using AccessorSubscript =
       typename AccessorCommonT::template AccessorSubscript<
@@ -2485,7 +2504,8 @@ class __SYCL_SPECIAL_CLASS local_accessor_base :
   local_accessor_base(handler &, const detail::code_location CodeLoc =
                                      detail::code_location::current())
 #ifdef __SYCL_DEVICE_ONLY__
-      : impl(range<AdjustedDim>{1}){}
+      : impl(range<AdjustedDim>{1}) {
+  }
 #else
       : LocalAccessorBaseHost(range<3>{1, 1, 1}, AdjustedDim, sizeof(DataT)) {
     detail::constructorNotification(nullptr, LocalAccessorBaseHost::impl.get(),
@@ -2494,11 +2514,10 @@ class __SYCL_SPECIAL_CLASS local_accessor_base :
   }
 #endif
 
-        template <int Dims = Dimensions,
-                  typename = detail::enable_if_t<Dims == 0>>
-        local_accessor_base(handler &, const property_list &propList,
-                            const detail::code_location CodeLoc =
-                                detail::code_location::current())
+  template <int Dims = Dimensions, typename = detail::enable_if_t<Dims == 0>>
+  local_accessor_base(
+      handler &, const property_list &propList,
+      const detail::code_location CodeLoc = detail::code_location::current())
 #ifdef __SYCL_DEVICE_ONLY__
       : impl(range<AdjustedDim>{1}) {
     (void)propList;
@@ -2517,7 +2536,8 @@ class __SYCL_SPECIAL_CLASS local_accessor_base :
       range<Dimensions> AllocationSize, handler &,
       const detail::code_location CodeLoc = detail::code_location::current())
 #ifdef __SYCL_DEVICE_ONLY__
-      : impl(AllocationSize){}
+      : impl(AllocationSize) {
+  }
 #else
       : LocalAccessorBaseHost(detail::convertToArrayOfN<3, 1>(AllocationSize),
                               AdjustedDim, sizeof(DataT)) {
@@ -2527,12 +2547,11 @@ class __SYCL_SPECIAL_CLASS local_accessor_base :
   }
 #endif
 
-        template <int Dims = Dimensions,
-                  typename = detail::enable_if_t<(Dims > 0)>>
-        local_accessor_base(range<Dimensions> AllocationSize, handler &,
-                            const property_list &propList,
-                            const detail::code_location CodeLoc =
-                                detail::code_location::current())
+  template <int Dims = Dimensions, typename = detail::enable_if_t<(Dims > 0)>>
+  local_accessor_base(
+      range<Dimensions> AllocationSize, handler &,
+      const property_list &propList,
+      const detail::code_location CodeLoc = detail::code_location::current())
 #ifdef __SYCL_DEVICE_ONLY__
       : impl(AllocationSize) {
     (void)propList;
@@ -2634,6 +2653,10 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS accessor<
   using local_acc =
       local_accessor_base<DataT, Dimensions, AccessMode, IsPlaceholder>;
 
+  static_assert(
+      !local_acc::IsConst || local_acc::IsAccessReadOnly,
+      "A const qualified DataT is only allowed for a read-only accessor");
+
   // Use base classes constructors
   using local_acc::local_acc;
 
@@ -2663,14 +2686,20 @@ class __SYCL_EBO __SYCL_SPECIAL_CLASS accessor<
 
 template <typename DataT, int Dimensions = 1>
 class __SYCL_EBO __SYCL_SPECIAL_CLASS __SYCL_TYPE(local_accessor) local_accessor
-    : public local_accessor_base<DataT, Dimensions, access::mode::read_write,
+    : public local_accessor_base<DataT, Dimensions,
+                                 detail::accessModeFromConstness<DataT>(),
                                  access::placeholder::false_t>,
       public detail::OwnerLessBase<local_accessor<DataT, Dimensions>> {
 
   using local_acc =
-      local_accessor_base<DataT, Dimensions, access::mode::read_write,
+      local_accessor_base<DataT, Dimensions,
+                          detail::accessModeFromConstness<DataT>(),
                           access::placeholder::false_t>;
 
+  static_assert(
+      !local_acc::IsConst || local_acc::IsAccessReadOnly,
+      "A const qualified DataT is only allowed for a read-only accessor");
+
   // Use base classes constructors
   using local_acc::local_acc;
 
diff --git a/sycl/test/basic_tests/accessor/accessor_ptr_alias.cpp b/sycl/test/basic_tests/accessor/accessor_ptr_alias.cpp
@@ -40,9 +40,11 @@ template <typename DataT, int Dims> void CheckLocalAccessor() {
 
 template <typename DataT, int Dims> void CheckAccessorForModes() {
   CheckDeviceAccessor<DataT, Dims, sycl::access::mode::read>();
-  CheckDeviceAccessor<DataT, Dims, sycl::access::mode::read_write>();
-  CheckDeviceAccessor<DataT, Dims, sycl::access::mode::write>();
   CheckLocalAccessor<DataT, Dims>();
+  if constexpr (!std::is_const<DataT>::value) {
+    CheckDeviceAccessor<DataT, Dims, sycl::access::mode::read_write>();
+    CheckDeviceAccessor<DataT, Dims, sycl::access::mode::write>();
+  }
 }
 
 template <typename DataT> void CheckAccessorForAllDimsAndModes() {
diff --git a/sycl/test/basic_tests/accessor/const-type-non-readonly-accessor.cpp b/sycl/test/basic_tests/accessor/const-type-non-readonly-accessor.cpp
@@ -0,0 +1,25 @@
+// RUN: %clangxx -fsycl -fsyntax-only -Xclang -verify -Xclang -verify-ignore-unexpected=note %s
+
+#include <sycl/sycl.hpp>
+
+using namespace sycl;
+
+constexpr size_t dataSize = 1;
+
+int main() {
+  const int a[dataSize] = {1};
+
+  try {
+    auto defaultQueue = queue{};
+    auto bufA = buffer<const int, 1>{a, range{dataSize}};
+    defaultQueue.submit([&](handler &cgh) {
+      // expected-error@sycl/accessor.hpp:* {{A const qualified DataT is only allowed for a read-only accessor}}
+      accessor accA{bufA, cgh, read_write};
+    });
+
+    defaultQueue.throw_asynchronous();
+  } catch (const exception &e) {
+    std::cout << "Exception caught: " << e.what() << std::endl;
+  }
+  return 0;
+}
diff --git a/sycl/test/basic_tests/accessor/host_accessor_get_pointer_type.cpp b/sycl/test/basic_tests/accessor/host_accessor_get_pointer_type.cpp
@@ -18,8 +18,10 @@ void CheckHostAccessor() {
 
 template <typename DataT, int Dims> void CheckHostAccessorForModes() {
   CheckHostAccessor<DataT, Dims, sycl::access::mode::read>();
-  CheckHostAccessor<DataT, Dims, sycl::access::mode::read_write>();
-  CheckHostAccessor<DataT, Dims, sycl::access::mode::write>();
+  if constexpr (!std::is_const<DataT>::value) {
+    CheckHostAccessor<DataT, Dims, sycl::access::mode::read_write>();
+    CheckHostAccessor<DataT, Dims, sycl::access::mode::write>();
+  }
 }
 
 template <typename DataT> void CheckHostAccessorForAllDimsAndModes() {
diff --git a/sycl/test/regression/accessor_subscript_and_ref_type.cpp b/sycl/test/regression/accessor_subscript_and_ref_type.cpp
@@ -81,13 +81,14 @@ template <typename DataT, access::mode AccessMode> void CheckAccAllDims() {
 }
 
 template <typename DataT> void CheckAccAllAccessModesAndDims() {
-  CheckAccAllDims<DataT, access::mode::write>();
-  CheckAccAllDims<DataT, access::mode::read_write>();
-  CheckAccAllDims<DataT, access::mode::discard_write>();
-  CheckAccAllDims<DataT, access::mode::discard_read_write>();
   CheckAccAllDims<DataT, access::mode::read>();
-  if constexpr (!std::is_const_v<DataT>)
+  if constexpr (!std::is_const_v<DataT>) {
+    CheckAccAllDims<DataT, access::mode::write>();
+    CheckAccAllDims<DataT, access::mode::read_write>();
+    CheckAccAllDims<DataT, access::mode::discard_write>();
+    CheckAccAllDims<DataT, access::mode::discard_read_write>();
     CheckAccAllDims<DataT, access::mode::atomic>();
+  }
 }
 
 int main() {
