From c1e1a52ef666c5c026709597cee6140fd6edb1dd Mon Sep 17 00:00:00 2001 From: Michael Schellenberger Costa Date: Thu, 8 Dec 2022 14:01:31 +0100 Subject: [PATCH] Improve constraints on `forward_property` to require a `upstream_resource` method --- .../get_property/forward_property.pass.cpp | 159 ++++++++++++------ include/cuda/memory_resource | 27 ++- 2 files changed, 128 insertions(+), 58 deletions(-) diff --git a/.upstream-tests/test/cuda/memory_resource/get_property/forward_property.pass.cpp b/.upstream-tests/test/cuda/memory_resource/get_property/forward_property.pass.cpp index c6f8597f19..fd7dcc167e 100644 --- a/.upstream-tests/test/cuda/memory_resource/get_property/forward_property.pass.cpp +++ b/.upstream-tests/test/cuda/memory_resource/get_property/forward_property.pass.cpp @@ -13,63 +13,114 @@ #include #include -struct prop_with_value { - using value_type = int; -}; -struct prop {}; - -template -struct derived_plain : public cuda::forward_property, Upstream> -{ - __host__ __device__ constexpr Upstream upstream_resource() const noexcept { return Upstream{}; } -}; - -struct upstream_with_valueless_property { - __host__ __device__ friend constexpr void get_property(const upstream_with_valueless_property&, prop) {} -}; -static_assert( cuda::has_property, prop>, ""); -static_assert(!cuda::has_property, prop_with_value>, ""); - -struct upstream_with_stateful_property { - __host__ __device__ friend constexpr int get_property(const upstream_with_stateful_property&, prop_with_value) { - return 42; - } -}; -static_assert(!cuda::has_property, prop>, ""); -static_assert( cuda::has_property, prop_with_value>, ""); - -struct upstream_with_both_properties { - __host__ __device__ friend constexpr void get_property(const upstream_with_both_properties&, prop) {} - __host__ __device__ friend constexpr int get_property(const upstream_with_both_properties&, prop_with_value) { - return 42; - } -}; -static_assert( cuda::has_property, prop>, ""); -static_assert( cuda::has_property, prop_with_value>, ""); - -struct derived_override : public cuda::forward_property -{ - __host__ __device__ constexpr upstream_with_both_properties upstream_resource() const noexcept { - return upstream_with_both_properties{}; - } - __host__ __device__ friend constexpr int get_property(const derived_override&, prop_with_value) { - return 1337; - } -}; +namespace has_upstream_resource { + struct Upstream{}; -__host__ __device__ constexpr bool test_stateful() { - using derived_no_override = derived_plain; - const derived_no_override without_override{}; - assert(get_property(without_override, prop_with_value{}) == 42); - - const derived_override with_override{}; - assert(get_property(with_override, prop_with_value{}) == 1337); + __device__ Upstream upstream{}; + + struct with_reference { + Upstream& upstream_resource() const { return upstream; } + }; + static_assert(cuda::__has_upstream_resource, ""); + + struct with_const_reference { + const Upstream& upstream_resource() const { return upstream; } + }; + static_assert(cuda::__has_upstream_resource, ""); + + struct with_value { + Upstream upstream_resource() const { return Upstream{}; } + }; + static_assert(cuda::__has_upstream_resource, ""); + + struct with_const_value { + const Upstream upstream_resource() const { return Upstream{}; } + }; + static_assert(cuda::__has_upstream_resource, ""); + + struct Convertible { + operator Upstream() { return Upstream{}; } + }; + + struct with_conversion { + Convertible upstream_resource() const { return Convertible{}; } + }; + static_assert(!cuda::__has_upstream_resource, ""); +} // namespace has_upstream_resource + +namespace forward_property { + struct prop_with_value { using value_type = int; }; + struct prop {}; + + template + struct derived_plain : public cuda::forward_property, Upstream> + { + constexpr Upstream upstream_resource() const noexcept { return Upstream{}; } + }; + + struct upstream_with_valueless_property { + friend constexpr void get_property(const upstream_with_valueless_property&, prop) {} + }; + static_assert( cuda::has_property, prop>, ""); + static_assert(!cuda::has_property, prop_with_value>, ""); + + struct upstream_with_stateful_property { + friend constexpr int get_property(const upstream_with_stateful_property&, prop_with_value) { + return 42; + } + }; + static_assert(!cuda::has_property, prop>, ""); + static_assert( cuda::has_property, prop_with_value>, ""); + + struct upstream_with_both_properties { + friend constexpr void get_property(const upstream_with_both_properties&, prop) {} + friend constexpr int get_property(const upstream_with_both_properties&, prop_with_value) { + return 42; + } + }; + static_assert( cuda::has_property, prop>, ""); + static_assert( cuda::has_property, prop_with_value>, ""); + + struct derived_override : public cuda::forward_property + { + constexpr upstream_with_both_properties upstream_resource() const noexcept { + return upstream_with_both_properties{}; + } + // Get called directly so needs to be annotated + __host__ __device__ friend constexpr int get_property(const derived_override&, prop_with_value) { + return 1337; + } + }; + + struct convertible_to_upstream { + operator upstream_with_both_properties() const noexcept { + return upstream_with_both_properties{}; + } + }; + + struct derived_with_converstin_upstream_resource : public cuda::forward_property + { + constexpr convertible_to_upstream upstream_resource() const noexcept { + return convertible_to_upstream{}; + } + }; + static_assert(!cuda::has_property); + + __host__ __device__ constexpr bool test_stateful() { + using derived_no_override = derived_plain; + const derived_no_override without_override{}; + assert(get_property(without_override, prop_with_value{}) == 42); + + const derived_override with_override{}; + assert(get_property(with_override, prop_with_value{}) == 1337); + + return true; + } +} // namespace forward_property - return true; -} int main(int, char**) { - test_stateful(); - static_assert(test_stateful(), ""); + forward_property::test_stateful(); + static_assert(forward_property::test_stateful(), ""); return 0; } diff --git a/include/cuda/memory_resource b/include/cuda/memory_resource index 42ae360a7c..31a869ef6b 100644 --- a/include/cuda/memory_resource +++ b/include/cuda/memory_resource @@ -130,17 +130,36 @@ _LIBCUDACXX_CONCEPT_FRAGMENT( template _LIBCUDACXX_CONCEPT has_property_with = _LIBCUDACXX_FRAGMENT(__has_property_with_, _Resource, _Property, _Return); +/// \concept __has_upstream_resource +/// \brief The \c __has_upstream_resource concept +template +_LIBCUDACXX_CONCEPT_FRAGMENT( + __has_upstream_resource_, + requires(const _Resource& __res)( + requires(_CUDA_VSTD::same_as<_CUDA_VSTD::__remove_const_ref_t, _Upstream>) + )); +template +_LIBCUDACXX_CONCEPT __has_upstream_resource = _LIBCUDACXX_FRAGMENT(__has_upstream_resource_, _Resource, _Upstream); + /// class forward_property -/// \brief The \c forward_property crtp temaplate simplifies the user facing side of forwarding properties -/// We can always tell people to just derive from it to properly forward all properties +/// \brief The \c forward_property crtp template simplifies the user facing side of forwarding properties +/// We can just derive from it to properly forward all properties template struct forward_property { + #if defined(__CUDACC__) && !defined(__NVCOMPILER) + #pragma nv_exec_check_disable + #endif _LIBCUDACXX_TEMPLATE(class _Property) (requires (!property_with_value<_Property>) _LIBCUDACXX_AND has_property<_Upstream, _Property>) _LIBCUDACXX_INLINE_VISIBILITY friend constexpr void get_property(const _Derived&, _Property) noexcept {} - _LIBCUDACXX_TEMPLATE(class _Property) - (requires (property_with_value<_Property>) _LIBCUDACXX_AND has_property<_Upstream, _Property>) + // The indirection is needed, otherwise the compiler might believe that _Derived is an incomplete type + #if defined(__CUDACC__) && !defined(__NVCOMPILER) + #pragma nv_exec_check_disable + #endif + _LIBCUDACXX_TEMPLATE(class _Property, class _Derived2 = _Derived) + (requires property_with_value<_Property> _LIBCUDACXX_AND has_property<_Upstream, _Property> _LIBCUDACXX_AND + __has_upstream_resource<_Derived2, _Upstream>) _LIBCUDACXX_INLINE_VISIBILITY friend constexpr __property_value_t<_Property> get_property( const _Derived& __res, _Property __prop) { return get_property(__res.upstream_resource(), __prop);