More review comment fixes

kokkos · Oct 3, 2024 · f86bfe8 · f86bfe8
1 parent e818f4d
commit f86bfe8
Show file tree

Hide file tree

Showing 3 changed files with 20 additions and 37 deletions.
diff --git a/include/experimental/__p0009_bits/utility.hpp b/include/experimental/__p0009_bits/utility.hpp
@@ -66,34 +66,34 @@ constexpr struct
   }
 } stride;
 
-template<class T>
-MDSPAN_INLINE_FUNCTION
-constexpr void maybe_unused_variable(const T&) {}
-
-
 // same as std::integral_constant but with __host__ __device__ annotations on
 // the implicit conversion function and the call operator
 template <class T, T v>
 struct integral_constant {
   using value_type         = T;
   using type               = integral_constant<T, v>;
 
+  static constexpr T value = v;
+
   MDSPAN_INLINE_FUNCTION_DEFAULTED
   constexpr integral_constant() = default;
-
-  MDSPAN_INLINE_FUNCTION_DEFAULTED
+
+  // These interop functions work, because other than the value_type operator
+  // everything of std::integral_constant works on device (defaulted functions)  
+  MDSPAN_FUNCTION
   constexpr integral_constant(std::integral_constant<T,v>) {};
+
+  MDSPAN_FUNCTION constexpr operator std::integral_constant<T,v>() const noexcept {
+    return std::integral_constant<T,v>{};
+  }
 
-  static constexpr T value = v;
-  MDSPAN_INLINE_FUNCTION constexpr operator value_type() const noexcept {
+  MDSPAN_FUNCTION constexpr operator value_type() const noexcept {
     return value;
   }
-  MDSPAN_INLINE_FUNCTION constexpr value_type operator()() const noexcept {
+
+  MDSPAN_FUNCTION constexpr value_type operator()() const noexcept {
     return value;
   }
-  MDSPAN_INLINE_FUNCTION constexpr operator std::integral_constant<T,v>() const noexcept {
-    return std::integral_constant<T,v>{};
-  }
 };
 
 // The tuple implementation only comes in play when using capabilities
@@ -162,16 +162,6 @@ constexpr const auto& get(const tuple<Args...>& vals) { return vals.template get
 template<class ... Elements>
 tuple(Elements ...) -> tuple<Elements...>;
 #endif
-
-template<class T, size_t ... Idx>
-constexpr auto c_array_to_std(std::index_sequence<Idx...>, const T(&values)[sizeof...(Idx)]) {
-  return std::array<T, sizeof...(Idx)>{values[Idx]...};
-}
-template<class T, size_t N>
-constexpr auto c_array_to_std(const T(&values)[N]) {
-  return c_array_to_std(std::make_index_sequence<N>(), values);
-}
-
 } // namespace detail
 
 constexpr struct mdspan_non_standard_tag {

diff --git a/include/experimental/__p2630_bits/submdspan_mapping.hpp b/include/experimental/__p2630_bits/submdspan_mapping.hpp
@@ -88,14 +88,7 @@ any_slice_out_of_bounds(const extents<IndexType, Exts...> &exts,
 template<class T, size_t N>
 struct sub_strides
 {
-  T values[N];
-};
-
-// Specialization to avoid 0-length arrays
-template<class T>
-struct sub_strides<T, 0>
-{
-  T values[1];
+  T values[N > 0 ? N : 1];
 };
 
 template <class SrcMapping, class... slice_strides, size_t... InvMapIdxs>

diff --git a/include/experimental/__p2642_bits/layout_padded.hpp b/include/experimental/__p2642_bits/layout_padded.hpp
@@ -358,15 +358,15 @@ class layout_left_padded<PaddingValue>::mapping {
     return exts;
   }
 
-  MDSPAN_INLINE_FUNCTION constexpr std::array<index_type, extents_type::rank()>
+  constexpr std::array<index_type, extents_type::rank()>
   strides() const noexcept {
     if constexpr (extents_type::rank() == 0) {
       return {};
     } else if constexpr (extents_type::rank() == 1) {
       return {1};
     } else {
       index_type value = 1;
-      index_type s[extents_type::rank()];
+      std::array<index_type, extents_type::rank()> s{};
       s[extent_to_pad_idx] = value;
       value *= padded_stride.value(0);
       for (rank_type r = extent_to_pad_idx + 1; r < extents_type::rank() - 1;
@@ -375,7 +375,7 @@ class layout_left_padded<PaddingValue>::mapping {
         value *= exts.extent(r);
       }
       s[extents_type::rank() - 1] = value;
-      return MDSPAN_IMPL_STANDARD_NAMESPACE::detail::c_array_to_std(s);
+      return s;
     }
   }
 
@@ -718,23 +718,23 @@ class layout_right_padded<PaddingValue>::mapping {
     return exts;
   }
 
-  MDSPAN_INLINE_FUNCTION constexpr std::array<index_type, extents_type::rank()>
+  constexpr std::array<index_type, extents_type::rank()>
   strides() const noexcept {
     if constexpr (extents_type::rank() == 0) {
       return {};
     } else if constexpr (extents_type::rank() == 1) {
       return {1};
     } else {
       index_type value = 1;
-      index_type s[extents_type::rank()];
+      std::array<index_type, extents_type::rank()> s{};
       s[extent_to_pad_idx] = value;
       value *= padded_stride.value(0);
       for (rank_type r = extent_to_pad_idx - 1; r > 0; --r) {
         s[r] = value;
         value *= exts.extent(r);
       }
       s[0] = value;
-      return MDSPAN_IMPL_STANDARD_NAMESPACE::detail::c_array_to_std(s);
+      return s;
     }
   }