Move std::remove_copy_if to its own file

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

@@ -30,6 +30,7 @@ set(files
   __algorithm/move_backward.h
   __algorithm/move.h
   __algorithm/none_of.h
+  __algorithm/remove_copy_if.h
   __algorithm/remove_copy.h
   __algorithm/remove_if.h
   __algorithm/remove.h

libcudacxx/include/cuda/std/detail/libcxx/include/__algorithm/remove_copy_if.h

@@ -0,0 +1,44 @@
+//===----------------------------------------------------------------------===//
+//
+// 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___ALGORITHM_REMOVE_COPY_IF_H
+#define _LIBCUDACXX___ALGORITHM_REMOVE_COPY_IF_H
+
+#ifndef __cuda_std__
+#  include <__config>
+#endif // __cuda_std__
+
+#if defined(_CCCL_IMPLICIT_SYSTEM_HEADER_GCC)
+#  pragma GCC system_header
+#elif defined(_CCCL_IMPLICIT_SYSTEM_HEADER_CLANG)
+#  pragma clang system_header
+#elif defined(_CCCL_IMPLICIT_SYSTEM_HEADER_MSVC)
+#  pragma system_header
+#endif // no system header
+
+_LIBCUDACXX_BEGIN_NAMESPACE_STD
+
+template <class _InputIterator, class _OutputIterator, class _Predicate>
+inline _LIBCUDACXX_HIDE_FROM_ABI _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 _OutputIterator
+remove_copy_if(_InputIterator __first, _InputIterator __last, _OutputIterator __result, _Predicate __pred)
+{
+  for (; __first != __last; ++__first)
+  {
+    if (!__pred(*__first))
+    {
+      *__result = *__first;
+      ++__result;
+    }
+  }
+  return __result;
+}
+
+_LIBCUDACXX_END_NAMESPACE_STD
+
+#endif // _LIBCUDACXX___ALGORITHM_REMOVE_COPY_IF_H

libcudacxx/include/cuda/std/detail/libcxx/include/algorithm

@@ -670,6 +670,7 @@ template <class BidirectionalIterator, class Compare>
 #include "__algorithm/move_backward.h"
 #include "__algorithm/move.h"
 #include "__algorithm/none_of.h"
+#include "__algorithm/remove_copy_if.h"
 #include "__algorithm/remove_copy.h"
 #include "__algorithm/remove_if.h"
 #include "__algorithm/remove.h"
@@ -740,24 +741,6 @@ public:
     bool operator()(const _T1& __x, const _T2& __y) {return __p_(__y, __x);}
 };
 
-// remove_copy_if
-
-template <class _InputIterator, class _OutputIterator, class _Predicate>
-inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX17
-_OutputIterator
-remove_copy_if(_InputIterator __first, _InputIterator __last, _OutputIterator __result, _Predicate __pred)
-{
-    for (; __first != __last; ++__first)
-    {
-        if (!__pred(*__first))
-        {
-            *__result = *__first;
-            ++__result;
-        }
-    }
-    return __result;
-}
-
 // unique
 
 template <class _ForwardIterator, class _BinaryPredicate>

libcudacxx/test/libcudacxx/std/algorithms/alg.modifying/alg.remove/remove_copy_if.pass.cpp

@@ -0,0 +1,86 @@
+//===----------------------------------------------------------------------===//
+//
+// 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
+//
+//===----------------------------------------------------------------------===//
+
+// <algorithm>
+
+// template<InputIterator InIter, OutputIterator<auto, InIter::reference> OutIter,
+//          Predicate<auto, InIter::value_type> Pred>
+//   requires CopyConstructible<Pred>
+//   constexpr OutIter         // constexpr after C++17
+//   remove_copy_if(InIter first, InIter last, OutIter result, Pred pred);
+
+#include <cuda/std/__algorithm>
+#include <cuda/std/functional>
+#include <cuda/std/cassert>
+
+#include "test_macros.h"
+#include "test_iterators.h"
+
+__host__ __device__ TEST_CONSTEXPR bool equalToTwo(const int v) noexcept { return v == 2; }
+
+template <class InIter, class OutIter>
+TEST_CONSTEXPR_CXX14 __host__ __device__ void test() {
+      constexpr int N = 9;
+    int ia[N] = {0, 1, 2, 3, 4, 2, 3, 4, 2};
+    constexpr int expected[N - 3] = {0, 1, 3, 4, 3, 4};
+    int ib[N] = {0};
+    OutIter r =
+        cuda::std::remove_copy_if(InIter(ia), InIter(ia + N), OutIter(ib), equalToTwo);
+    assert(base(r) == ib + N - 3);
+    for (int i = 0; i < N - 3; ++i) {
+        assert(ib[i] == expected[i]);
+    }
+
+    for (int i = N - 3; i < N; ++i) {
+        assert(ib[i] == 0);
+    }
+}
+
+TEST_CONSTEXPR_CXX14 __host__ __device__ bool test() {
+  test<cpp17_input_iterator<const int*>, cpp17_output_iterator<int*> >();
+  test<cpp17_input_iterator<const int*>, forward_iterator<int*> >();
+  test<cpp17_input_iterator<const int*>, bidirectional_iterator<int*> >();
+  test<cpp17_input_iterator<const int*>, random_access_iterator<int*> >();
+  test<cpp17_input_iterator<const int*>, int*>();
+
+  test<forward_iterator<const int*>, cpp17_output_iterator<int*> >();
+  test<forward_iterator<const int*>, forward_iterator<int*> >();
+  test<forward_iterator<const int*>, bidirectional_iterator<int*> >();
+  test<forward_iterator<const int*>, random_access_iterator<int*> >();
+  test<forward_iterator<const int*>, int*>();
+
+  test<bidirectional_iterator<const int*>, cpp17_output_iterator<int*> >();
+  test<bidirectional_iterator<const int*>, forward_iterator<int*> >();
+  test<bidirectional_iterator<const int*>, bidirectional_iterator<int*> >();
+  test<bidirectional_iterator<const int*>, random_access_iterator<int*> >();
+  test<bidirectional_iterator<const int*>, int*>();
+
+  test<random_access_iterator<const int*>, cpp17_output_iterator<int*> >();
+  test<random_access_iterator<const int*>, forward_iterator<int*> >();
+  test<random_access_iterator<const int*>, bidirectional_iterator<int*> >();
+  test<random_access_iterator<const int*>, random_access_iterator<int*> >();
+  test<random_access_iterator<const int*>, int*>();
+
+  test<const int*, cpp17_output_iterator<int*> >();
+  test<const int*, forward_iterator<int*> >();
+  test<const int*, bidirectional_iterator<int*> >();
+  test<const int*, random_access_iterator<int*> >();
+  test<const int*, int*>();
+
+  return true;
+}
+
+int main(int, char**) {
+  test();
+
+#if TEST_STD_VER >= 14
+  static_assert(test(), "");
+#endif // TEST_STD_VER >= 14
+
+  return 0;
+}