diff --git a/include/experimental/__p0009_bits/utility.hpp b/include/experimental/__p0009_bits/utility.hpp index 7dd2f26b..54b63cf3 100644 --- a/include/experimental/__p0009_bits/utility.hpp +++ b/include/experimental/__p0009_bits/utility.hpp @@ -66,11 +66,6 @@ constexpr struct } } stride; -template -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 @@ -78,22 +73,27 @@ struct integral_constant { using value_type = T; using type = integral_constant; + 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) {}; + + MDSPAN_FUNCTION constexpr operator std::integral_constant() const noexcept { + return std::integral_constant{}; + } - 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() const noexcept { - return std::integral_constant{}; - } }; // The tuple implementation only comes in play when using capabilities @@ -162,16 +162,6 @@ constexpr const auto& get(const tuple& vals) { return vals.template get template tuple(Elements ...) -> tuple; #endif - -template -constexpr auto c_array_to_std(std::index_sequence, const T(&values)[sizeof...(Idx)]) { - return std::array{values[Idx]...}; -} -template -constexpr auto c_array_to_std(const T(&values)[N]) { - return c_array_to_std(std::make_index_sequence(), 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 index 52dfd77b..2a2cdf76 100644 --- 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 &exts, template struct sub_strides { - T values[N]; -}; - -// Specialization to avoid 0-length arrays -template -struct sub_strides -{ - T values[1]; + T values[N > 0 ? N : 1]; }; template diff --git a/include/experimental/__p2642_bits/layout_padded.hpp b/include/experimental/__p2642_bits/layout_padded.hpp index 459f5b19..a714090e 100644 --- a/include/experimental/__p2642_bits/layout_padded.hpp +++ b/include/experimental/__p2642_bits/layout_padded.hpp @@ -358,7 +358,7 @@ class layout_left_padded::mapping { return exts; } - MDSPAN_INLINE_FUNCTION constexpr std::array + constexpr std::array strides() const noexcept { if constexpr (extents_type::rank() == 0) { return {}; @@ -366,7 +366,7 @@ class layout_left_padded::mapping { return {1}; } else { index_type value = 1; - index_type s[extents_type::rank()]; + std::array 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::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,7 +718,7 @@ class layout_right_padded::mapping { return exts; } - MDSPAN_INLINE_FUNCTION constexpr std::array + constexpr std::array strides() const noexcept { if constexpr (extents_type::rank() == 0) { return {}; @@ -726,7 +726,7 @@ class layout_right_padded::mapping { return {1}; } else { index_type value = 1; - index_type s[extents_type::rank()]; + std::array s{}; s[extent_to_pad_idx] = value; value *= padded_stride.value(0); for (rank_type r = extent_to_pad_idx - 1; r > 0; --r) { @@ -734,7 +734,7 @@ class layout_right_padded::mapping { value *= exts.extent(r); } s[0] = value; - return MDSPAN_IMPL_STANDARD_NAMESPACE::detail::c_array_to_std(s); + return s; } }