[FEA] Implement iter_move CPO

libcudacxx/.upstream-tests/test/std/iterators/iterator.requirements/iterator.cust/iterator.cust.move/iter_move.nodiscard.verify.cpp

@@ -0,0 +1,37 @@
+//===----------------------------------------------------------------------===//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+// SPDX-FileCopyrightText: Copyright (c) 2023 NVIDIA CORPORATION & AFFILIATES.
+//
+//===----------------------------------------------------------------------===//
+
+// UNSUPPORTED: c++03, c++11, c++14
+// UNSUPPORTED: LIBCUDACXX-has-no-incomplete-ranges
+
+// Test the [[nodiscard]] extension in libc++.
+
+// template<class I>
+// unspecified iter_move;
+
+#include <cuda/std/iterator>
+
+struct WithADL {
+  WithADL() = default;
+  __host__ __device__ constexpr decltype(auto) operator*() const noexcept;
+  __host__ __device__ constexpr WithADL& operator++() noexcept;
+  __host__ __device__ constexpr void operator++(int) noexcept;
+  __host__ __device__ constexpr bool operator==(WithADL const&) const noexcept;
+  __host__ __device__ friend constexpr auto iter_move(WithADL&) { return 0; }
+};
+
+int main(int, char**) {
+  int* noADL = nullptr;
+  cuda::std::ranges::iter_move(noADL); // expected-warning {{ignoring return value of function declared with 'nodiscard' attribute}}
+
+  WithADL adl;
+  cuda::std::ranges::iter_move(adl); // expected-warning {{ignoring return value of function declared with 'nodiscard' attribute}}
+
+  return 0;
+}

libcudacxx/.upstream-tests/test/std/iterators/iterator.requirements/iterator.cust/iterator.cust.move/iter_move.pass.cpp

@@ -0,0 +1,222 @@
+//===----------------------------------------------------------------------===//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+// SPDX-FileCopyrightText: Copyright (c) 2023 NVIDIA CORPORATION & AFFILIATES.
+//
+//===----------------------------------------------------------------------===//
+
+// UNSUPPORTED: c++03, c++11, c++14
+
+// template<class I>
+// unspecified iter_move;
+
+#include <cuda/std/iterator>
+
+#include <cuda/std/array>
+#include <cuda/std/cassert>
+#include <cuda/std/utility>
+
+#include "test_macros.h"
+
+#include "../unqualified_lookup_wrapper.h"
+
+using IterMoveT = decltype(cuda::std::ranges::iter_move);
+
+// Wrapper around an iterator for testing `iter_move` when an unqualified call to `iter_move` isn't
+// possible.
+template <typename I>
+class iterator_wrapper {
+public:
+  iterator_wrapper() = default;
+
+  __host__ __device__ constexpr explicit iterator_wrapper(I i) noexcept : base_(cuda::std::move(i)) {}
+
+  // `noexcept(false)` is used to check that this operator is called.
+  __host__ __device__ constexpr decltype(auto) operator*() const& noexcept(false) { return *base_; }
+
+  // `noexcept` is used to check that this operator is called.
+  __host__ __device__ constexpr auto&& operator*() && noexcept { return cuda::std::move(*base_); }
+
+  __host__ __device__ constexpr iterator_wrapper& operator++() noexcept {
+    ++base_;
+    return *this;
+  }
+
+  __host__ __device__ constexpr void operator++(int) noexcept { ++base_; }
+
+  __host__ __device__ constexpr bool operator==(iterator_wrapper const& other) const noexcept { return base_ == other.base_; }
+#if TEST_STD_VER < 20
+  __host__ __device__ constexpr bool operator!=(iterator_wrapper const& other) const noexcept { return base_ != other.base_; }
+#endif
+
+private:
+  I base_ = I{};
+};
+
+template <typename It, typename Out>
+__host__ __device__ constexpr void unqualified_lookup_move(It first_, It last_, Out result_first_, Out result_last_) {
+  auto first = ::check_unqualified_lookup::unqualified_lookup_wrapper<It>{cuda::std::move(first_)};
+  auto last = ::check_unqualified_lookup::unqualified_lookup_wrapper<It>{cuda::std::move(last_)};
+  auto result_first = ::check_unqualified_lookup::unqualified_lookup_wrapper<It>{cuda::std::move(result_first_)};
+  auto result_last = ::check_unqualified_lookup::unqualified_lookup_wrapper<It>{cuda::std::move(result_last_)};
+
+  static_assert(!noexcept(cuda::std::ranges::iter_move(first)), "unqualified-lookup case not being chosen");
+
+  for (; first != last && result_first != result_last; (void)++first, ++result_first) {
+    *result_first = cuda::std::ranges::iter_move(first);
+  }
+}
+
+template <typename It, typename Out>
+__host__ __device__ constexpr void lvalue_move(It first_, It last_, Out result_first_, Out result_last_) {
+  auto first = iterator_wrapper<It>{cuda::std::move(first_)};
+  auto last = ::iterator_wrapper<It>{cuda::std::move(last_)};
+  auto result_first = iterator_wrapper<It>{cuda::std::move(result_first_)};
+  auto result_last = iterator_wrapper<It>{cuda::std::move(result_last_)};
+
+  static_assert(!noexcept(cuda::std::ranges::iter_move(first)), "`operator*() const&` is not noexcept, and there's no hidden "
+                                                          "friend iter_move.");
+
+  for (; first != last && result_first != result_last; (void)++first, ++result_first) {
+    *result_first = cuda::std::ranges::iter_move(first);
+  }
+}
+
+template <typename It, typename Out>
+__host__ __device__ constexpr void rvalue_move(It first_, It last_, Out result_first_, Out result_last_) {
+  auto first = iterator_wrapper<It>{cuda::std::move(first_)};
+  auto last = iterator_wrapper<It>{cuda::std::move(last_)};
+  auto result_first = iterator_wrapper<It>{cuda::std::move(result_first_)};
+  auto result_last = iterator_wrapper<It>{cuda::std::move(result_last_)};
+
+  static_assert(noexcept(cuda::std::ranges::iter_move(cuda::std::move(first))),
+                "`operator*() &&` is noexcept, and there's no hidden friend iter_move.");
+
+  for (; first != last && result_first != result_last; (void)++first, ++result_first) {
+    auto i = first;
+    *result_first = cuda::std::ranges::iter_move(cuda::std::move(i));
+  }
+}
+
+template <bool NoExcept>
+struct WithADL {
+  WithADL() = default;
+  __host__ __device__ constexpr int operator*() const { return 0; }
+  __host__ __device__ constexpr WithADL& operator++();
+  __host__ __device__ constexpr void operator++(int);
+  __host__ __device__ constexpr bool operator==(WithADL const&) const;
+  __host__ __device__ friend constexpr int iter_move(WithADL&&) noexcept(NoExcept) { return 0; }
+};
+
+template <bool NoExcept>
+struct WithoutADL {
+  WithoutADL() = default;
+  __host__ __device__ constexpr int operator*() const noexcept(NoExcept) { return 0; }
+  __host__ __device__ constexpr WithoutADL& operator++();
+  __host__ __device__ constexpr void operator++(int);
+  __host__ __device__ constexpr bool operator==(WithoutADL const&) const;
+};
+
+template <class It, class Pred>
+__host__ __device__ constexpr bool all_of(It first, It last, Pred pred) {
+    for (; first != last; ++first) {
+        if (!pred(*first)) {
+            return false;
+        }
+    }
+    return true;
+}
+
+__host__ __device__ constexpr bool test() {
+  constexpr int full_size = 100;
+  constexpr int half_size = full_size / 2;
+  constexpr int reset = 0;
+  move_tracker v1[full_size];
+
+  struct move_counter_is {
+    __host__ __device__ constexpr move_counter_is(const int counter) : _counter(counter) {}
+
+    __host__ __device__ constexpr bool operator()(move_tracker const& x) {
+      return x.moves() == _counter;
+    }
+
+    const int _counter;
+  };
+
+  move_tracker v2[half_size];
+  unqualified_lookup_move(cuda::std::begin(v1), cuda::std::end(v1), cuda::std::begin(v2), cuda::std::end(v2));
+  assert(all_of(cuda::std::cbegin(v1), cuda::std::cend(v1), move_counter_is(reset)));
+  assert(all_of(cuda::std::cbegin(v2), cuda::std::cend(v2), move_counter_is(1)));
+
+  move_tracker v3[half_size];
+  unqualified_lookup_move(cuda::std::begin(v1) + half_size, cuda::std::end(v1), cuda::std::begin(v3), cuda::std::end(v3));
+  assert(all_of(cuda::std::cbegin(v1), cuda::std::cend(v1), move_counter_is(reset)));
+  assert(all_of(cuda::std::cbegin(v3), cuda::std::cend(v3), move_counter_is(1)));
+
+  move_tracker v4[half_size];
+  unqualified_lookup_move(cuda::std::begin(v3), cuda::std::end(v3), cuda::std::begin(v4), cuda::std::end(v4));
+  assert(all_of(cuda::std::cbegin(v3), cuda::std::cend(v3), move_counter_is(reset)));
+  assert(all_of(cuda::std::cbegin(v4), cuda::std::cend(v4), move_counter_is(2)));
+
+  lvalue_move(cuda::std::begin(v2), cuda::std::end(v2), cuda::std::begin(v1) + half_size, cuda::std::end(v1));
+  assert(all_of(cuda::std::cbegin(v2), cuda::std::cend(v2), move_counter_is(reset)));
+  assert(all_of(cuda::std::cbegin(v1) + half_size, cuda::std::cend(v1), move_counter_is(2)));
+
+  lvalue_move(cuda::std::begin(v4), cuda::std::end(v4), cuda::std::begin(v1), cuda::std::end(v1));
+  assert(all_of(cuda::std::cbegin(v4), cuda::std::cend(v4), move_counter_is(reset)));
+  assert(all_of(cuda::std::cbegin(v1), cuda::std::cbegin(v1) + half_size, move_counter_is(3)));
+
+  rvalue_move(cuda::std::begin(v1), cuda::std::end(v1), cuda::std::begin(v2), cuda::std::end(v2));
+  assert(all_of(cuda::std::cbegin(v1), cuda::std::cbegin(v1) + half_size, move_counter_is(reset)));
+  assert(all_of(cuda::std::cbegin(v2), cuda::std::cend(v2), move_counter_is(4)));
+
+  rvalue_move(cuda::std::begin(v1) + half_size, cuda::std::end(v1), cuda::std::begin(v3), cuda::std::end(v3));
+  assert(all_of(cuda::std::cbegin(v1), cuda::std::cend(v1), move_counter_is(reset)));
+  assert(all_of(cuda::std::cbegin(v3), cuda::std::cend(v3), move_counter_is(3)));
+
+  auto unscoped = check_unqualified_lookup::unscoped_enum::a;
+  assert(cuda::std::ranges::iter_move(unscoped) == check_unqualified_lookup::unscoped_enum::a);
+  assert(!noexcept(cuda::std::ranges::iter_move(unscoped)));
+
+  auto scoped = check_unqualified_lookup::scoped_enum::a;
+  assert(cuda::std::ranges::iter_move(scoped) == nullptr);
+  assert(noexcept(cuda::std::ranges::iter_move(scoped)));
+
+  auto some_union = check_unqualified_lookup::some_union{0};
+  assert(cuda::std::ranges::iter_move(some_union) == 0);
+  assert(!noexcept(cuda::std::ranges::iter_move(some_union)));
+
+  // Check noexcept-correctness
+  static_assert(noexcept(cuda::std::ranges::iter_move(cuda::std::declval<WithADL<true>>())));
+  static_assert(noexcept(cuda::std::ranges::iter_move(cuda::std::declval<WithoutADL<true>>())));
+// old GCC seems to fall over the chaining of the noexcept clauses here
+#if (!defined(TEST_COMPILER_GCC) || __GNUC__ >= 9)
+  static_assert(!noexcept(cuda::std::ranges::iter_move(cuda::std::declval<WithADL<false>>())));
+  static_assert(!noexcept(cuda::std::ranges::iter_move(cuda::std::declval<WithoutADL<false>>())));
+#endif
+
+  return true;
+}
+
+#ifndef _LIBCUDACXX_COMPILER_NVCC_BELOW_11_3 // nvcc segfaults here
+static_assert(!cuda::std::is_invocable_v<IterMoveT, int*, int*>); // too many arguments
+static_assert(!cuda::std::is_invocable_v<IterMoveT, int>);
+#endif // _LIBCUDACXX_COMPILER_NVCC_BELOW_11_3
+
+#if TEST_STD_VER > 17
+// Test ADL-proofing.
+struct Incomplete;
+template<class T> struct Holder { T t; };
+static_assert(cuda::std::is_invocable_v<IterMoveT, Holder<Incomplete>**>);
+static_assert(cuda::std::is_invocable_v<IterMoveT, Holder<Incomplete>**&>);
+#endif
+
+int main(int, char**)
+{
+  test();
+  static_assert(test());
+
+  return 0;
+}

libcudacxx/.upstream-tests/test/std/iterators/iterator.requirements/iterator.cust/iterator.cust.move/iter_rvalue_reference_t.compile.pass.cpp

@@ -0,0 +1,30 @@
+//===----------------------------------------------------------------------===//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+// SPDX-FileCopyrightText: Copyright (c) 2023 NVIDIA CORPORATION & AFFILIATES.
+//
+//===----------------------------------------------------------------------===//
+
+// UNSUPPORTED: c++03, c++11, c++14
+
+// template<class I>
+// using iter_rvalue_reference;
+
+#include <cuda/std/iterator>
+
+static_assert(cuda::std::same_as<cuda::std::iter_rvalue_reference_t<int*>, int&&>);
+static_assert(cuda::std::same_as<cuda::std::iter_rvalue_reference_t<const int*>, const int&&>);
+
+__host__ __device__ void test_undefined_internal() {
+  struct A {
+    __host__ __device__ int& operator*() const;
+  };
+  static_assert(cuda::std::same_as<cuda::std::iter_rvalue_reference_t<A>, int&&>);
+}
+
+int main(int, char**)
+{
+  return 0;
+}

libcudacxx/.upstream-tests/test/std/iterators/iterator.requirements/iterator.cust/unqualified_lookup_wrapper.h

@@ -0,0 +1,87 @@
+//===----------------------------------------------------------------------===//
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+// SPDX-FileCopyrightText: Copyright (c) 2023 NVIDIA CORPORATION & AFFILIATES.
+//
+//===----------------------------------------------------------------------===//
+#ifndef LIBCUDACXX_TEST_STD_ITERATOR_UNQUALIFIED_LOOKUP_WRAPPER
+#define LIBCUDACXX_TEST_STD_ITERATOR_UNQUALIFIED_LOOKUP_WRAPPER
+
+#include <cuda/std/iterator>
+#include <cuda/std/utility>
+
+#include "test_macros.h"
+
+namespace check_unqualified_lookup {
+// Wrapper around an iterator for testing unqualified calls to `iter_move` and `iter_swap`.
+template <typename I>
+class unqualified_lookup_wrapper {
+public:
+  unqualified_lookup_wrapper() = default;
+
+  __host__ __device__ constexpr explicit unqualified_lookup_wrapper(I i) noexcept : base_(cuda::std::move(i)) {}
+
+  __host__ __device__ constexpr decltype(auto) operator*() const noexcept { return *base_; }
+
+  __host__ __device__ constexpr unqualified_lookup_wrapper& operator++() noexcept {
+    ++base_;
+    return *this;
+  }
+
+  __host__ __device__ constexpr void operator++(int) noexcept { ++base_; }
+
+  __host__ __device__ constexpr bool operator==(unqualified_lookup_wrapper const& other) const noexcept {
+    return base_ == other.base_;
+  }
+#if TEST_STD_VER < 20
+  __host__ __device__ constexpr bool operator!=(unqualified_lookup_wrapper const& other) const noexcept {
+    return base_ != other.base_;
+  }
+#endif
+
+  // Delegates `cuda::std::ranges::iter_move` for the underlying iterator. `noexcept(false)` will be used
+  // to ensure that the unqualified-lookup overload is chosen.
+  __host__ __device__ friend constexpr decltype(auto) iter_move(unqualified_lookup_wrapper& i) noexcept(false) {
+    return cuda::std::ranges::iter_move(i.base_);
+  }
+
+private:
+  I base_ = I{};
+};
+
+enum unscoped_enum { a, b, c };
+__host__ __device__ constexpr unscoped_enum iter_move(unscoped_enum& e) noexcept(false) { return e; }
+
+enum class scoped_enum { a, b, c };
+__host__ __device__ constexpr scoped_enum* iter_move(scoped_enum&) noexcept { return nullptr; }
+
+union some_union {
+  int x;
+  double y;
+};
+__host__ __device__ constexpr int iter_move(some_union& u) noexcept(false) { return u.x; }
+
+} // namespace check_unqualified_lookup
+
+class move_tracker {
+public:
+  move_tracker() = default;
+  __host__ __device__ constexpr move_tracker(move_tracker&& other) noexcept : moves_{other.moves_ + 1} { other.moves_ = 0; }
+  __host__ __device__ constexpr move_tracker& operator=(move_tracker&& other) noexcept {
+    moves_ = other.moves_ + 1;
+    other.moves_ = 0;
+    return *this;
+  }
+
+  move_tracker(move_tracker const& other) = delete;
+  move_tracker& operator=(move_tracker const& other) = delete;
+
+  __host__ __device__ constexpr int moves() const noexcept { return moves_; }
+
+private:
+  int moves_ = 0;
+};
+
+#endif // LIBCUDACXX_TEST_STD_ITERATOR_UNQUALIFIED_LOOKUP_WRAPPER

libcudacxx/include/cuda/std/detail/libcxx/include/CMakeLists.txt

@@ -93,6 +93,7 @@ set(files
   __iterator/incrementable_traits.h
   __iterator/istream_iterator.h
   __iterator/istreambuf_iterator.h
+  __iterator/iter_move.h
   __iterator/iterator.h
   __iterator/iterator_traits.h
   __iterator/move_iterator.h