[SYCL] Complete nd_item interface updates

Author: crystarm

sycl/include/sycl/nd_item.hpp

@@ -47,6 +47,8 @@ template <int Dimensions = 1> class nd_item {
 public:
   static constexpr int dimensions = Dimensions;
 
+  nd_item() = delete;
+
   id<Dimensions> get_global_id() const noexcept {
 #ifdef __SYCL_DEVICE_ONLY__
     return __spirv::initBuiltInGlobalInvocationId<Dimensions, id<Dimensions>>();
@@ -199,7 +201,7 @@ template <int Dimensions = 1> class nd_item {
 #ifndef __INTEL_PREVIEW_BREAKING_CHANGES
   __SYCL2020_DEPRECATED("use sycl::group_barrier() free function instead")
   void barrier([[maybe_unused]] access::fence_space accessSpace =
-                   access::fence_space::global_and_local) const noexcept {
+                   access::fence_space::global_and_local) const {
 #ifdef __SYCL_DEVICE_ONLY__
     uint32_t flags = _V1::detail::getSPIRVMemorySemanticsMask(accessSpace);
     __spirv_ControlBarrier(__spv::Scope::Workgroup, __spv::Scope::Workgroup,
@@ -217,7 +219,7 @@ template <int Dimensions = 1> class nd_item {
                                     accessMode == access::mode::write ||
                                     accessMode == access::mode::read_write,
                                 access::fence_space>
-          accessSpace = access::fence_space::global_and_local) const noexcept {
+          accessSpace = access::fence_space::global_and_local) const {
 #if __SYCL_DEVICE_ONLY__
     uint32_t flags = detail::getSPIRVMemorySemanticsMask(accessSpace);
     // TODO: currently, there is no good way in SPIR-V to set the memory
@@ -528,7 +530,6 @@ template <int Dimensions = 1> class nd_item {
 
 protected:
   friend class detail::Builder;
-  nd_item() {}
   nd_item(const item<Dimensions, true> &, const item<Dimensions, false> &,
           const group<Dimensions> &) {}
 

sycl/test/basic_tests/nd_item_interface.cpp

@@ -35,6 +35,8 @@ struct has_member_not_equal<
 
 using Item = sycl::nd_item<1>;
 
+static_assert(!std::is_default_constructible_v<Item>);
+
 static_assert(noexcept(std::declval<const Item &>().get_global_id()));
 static_assert(noexcept(std::declval<const Item &>().get_global_id(0)));
 static_assert(noexcept(std::declval<const Item &>().get_global_linear_id()));
@@ -54,11 +56,6 @@ static_assert(noexcept(std::declval<const Item &>().get_local_range(0)));
 static_assert(noexcept(std::declval<const Item &>().get_offset()));
 static_assert(noexcept(std::declval<const Item &>().get_nd_range()));
 
-#ifndef __INTEL_PREVIEW_BREAKING_CHANGES
-static_assert(noexcept(std::declval<const Item &>().barrier()));
-static_assert(noexcept(std::declval<const Item &>().mem_fence()));
-#endif
-
 static_assert(!has_member_equal<Item>::value);
 static_assert(!has_member_not_equal<Item>::value);
 static_assert(std::is_same_v<decltype(std::declval<const Item &>() ==