From b6bcef797fc0872ad7a27e928dc2f46ec8546424 Mon Sep 17 00:00:00 2001
From: Anton Reinhard <anton.reinhard@wandelbots.com>
Date: Wed, 2 Feb 2022 14:07:59 +0100
Subject: [PATCH] Cleanup, namespaces

---
 example/CMakeLists.txt                        |   1 -
 .../src/constantIterator-main.cpp             |   2 +-
 .../constantIteratorBenchmark/CMakeLists.txt  |   4 -
 .../src/constantIteratorBenchmark-main.cpp    | 208 ----------
 .../vikunja/mem/iterator/ConstantIterator.hpp | 374 +++++++++---------
 .../reduce/bench_vikunja_constiter_reduce.cpp |   2 +-
 test/integ/transform/src/ConstantIterator.cpp |   2 +-
 test/unit/iterator/src/ConstantIterator.cpp   |   2 +-
 8 files changed, 188 insertions(+), 407 deletions(-)
 delete mode 100644 example/constantIteratorBenchmark/CMakeLists.txt
 delete mode 100644 example/constantIteratorBenchmark/src/constantIteratorBenchmark-main.cpp

diff --git a/example/CMakeLists.txt b/example/CMakeLists.txt
index 8172f4c..f039244 100644
--- a/example/CMakeLists.txt
+++ b/example/CMakeLists.txt
@@ -1,5 +1,4 @@
 cmake_minimum_required(VERSION 3.18)
 add_subdirectory("constantIterator/")
-add_subdirectory("constantIteratorBenchmark/")
 add_subdirectory("reduce/")
 add_subdirectory("transform/")
diff --git a/example/constantIterator/src/constantIterator-main.cpp b/example/constantIterator/src/constantIterator-main.cpp
index 166908f..9dc7652 100644
--- a/example/constantIterator/src/constantIterator-main.cpp
+++ b/example/constantIterator/src/constantIterator-main.cpp
@@ -55,7 +55,7 @@ int main()
               << "Testing constant iterator with value: 10\n";
 
     // Create the constant iterator
-    vikunja::mem::iterator::ConstantIterator<TRed> constantIter(10);
+    vikunja::iterator::ConstantIterator<TRed> constantIter(10);
 
     // REDUCE CALL:
     // Takes the arguments: accelerator device, host device, accelerator queue, size of data, pointer-like to memory,
diff --git a/example/constantIteratorBenchmark/CMakeLists.txt b/example/constantIteratorBenchmark/CMakeLists.txt
deleted file mode 100644
index dceec79..0000000
--- a/example/constantIteratorBenchmark/CMakeLists.txt
+++ /dev/null
@@ -1,4 +0,0 @@
-cmake_minimum_required(VERSION 3.18)
-set(_TARGET_NAME "example_constant_iterator_benchmark")
-alpaka_add_executable(${_TARGET_NAME} src/constantIteratorBenchmark-main.cpp)
-target_link_libraries(${_TARGET_NAME} PUBLIC vikunja::internalvikunja)
diff --git a/example/constantIteratorBenchmark/src/constantIteratorBenchmark-main.cpp b/example/constantIteratorBenchmark/src/constantIteratorBenchmark-main.cpp
deleted file mode 100644
index ce1a3ba..0000000
--- a/example/constantIteratorBenchmark/src/constantIteratorBenchmark-main.cpp
+++ /dev/null
@@ -1,208 +0,0 @@
-/* Copyright 2021 Hauke Mewes, Simeon Ehrig, Victor
- *
- * This file is part of vikunja.
- *
- * This Source Code Form is subject to the terms of the Mozilla Public
- * License, v. 2.0. If a copy of the MPL was not distributed with this
- * file, You can obtain one at http://mozilla.org/MPL/2.0/.
- */
-
-#include <vikunja/mem/iterator/ConstantIterator.hpp>
-#include <vikunja/reduce/reduce.hpp>
-
-#include <alpaka/alpaka.hpp>
-
-#include <chrono>
-#include <iostream>
-
-int main()
-{
-// Define the accelerator here. Must be one of the enabled accelerators.
-#ifdef ALPAKA_ACC_GPU_CUDA_ENABLED
-    using TAcc = alpaka::AccGpuCudaRt<alpaka::DimInt<3u>, std::uint64_t>;
-#else
-    using TAcc = alpaka::AccCpuSerial<alpaka::DimInt<3u>, std::uint64_t>;
-#endif
-
-    // Type of the data that will be reduced
-    using TRed = uint64_t;
-
-    // Alpaka index type
-    using Idx = alpaka::Idx<TAcc>;
-    // Alpaka dimension type
-    using Dim = alpaka::Dim<TAcc>;
-    // Type of the extent vector
-    using Vec = alpaka::Vec<Dim, Idx>;
-    // Find the index of the CUDA blockIdx.x component. Alpaka somehow reverses
-    // these, i.e. the x component of cuda is always the last value in the vector
-    constexpr Idx xIndex = Dim::value - 1u;
-    // number of elements to reduce
-    const Idx n = static_cast<Idx>(1000000);
-    // create extent
-    Vec extent(Vec::all(static_cast<Idx>(1)));
-    extent[xIndex] = n;
-
-    // Value for constant iterator
-    const TRed constantIterVal = 10;
-    // Number of benchmark iteration
-    const int benchmarkIterations = 10000;
-
-    // define device, platform, and queue types.
-    using DevAcc = alpaka::Dev<TAcc>;
-    using PltfAcc = alpaka::Pltf<DevAcc>;
-    // using QueueAcc = alpaka::test::queue::DefaultQueue<alpaka::Dev<TAcc>>;
-    using PltfHost = alpaka::PltfCpu;
-    using DevHost = alpaka::Dev<PltfHost>;
-    using QueueAcc = alpaka::Queue<TAcc, alpaka::Blocking>;
-
-    // Get the host device.
-    DevHost devHost(alpaka::getDevByIdx<PltfHost>(0u));
-    // Select a device to execute on.
-    DevAcc devAcc(alpaka::getDevByIdx<PltfAcc>(0u));
-    // Get a queue on the accelerator device.
-    QueueAcc queueAcc(devAcc);
-
-    // allocate memory both on host and device.
-    auto deviceMem(alpaka::allocBuf<TRed, Idx>(devAcc, extent));
-    auto hostMem(alpaka::allocBuf<TRed, Idx>(devHost, extent));
-    // Fill memory on host with constantIterVal.
-    TRed* hostNative = alpaka::getPtrNative(hostMem);
-    for(Idx i = 0; i < n; ++i)
-    {
-        // std::cout << i << "\n";
-        hostNative[i] = static_cast<TRed>(constantIterVal);
-    }
-    // Copy to accelerator.
-    alpaka::memcpy(queueAcc, deviceMem, hostMem, extent);
-
-    // Use Lambda function for reduction
-    auto sum = [] ALPAKA_FN_HOST_ACC(TRed const i, TRed const j) { return i + j; };
-    auto doubleNum = [] ALPAKA_FN_HOST_ACC(TRed const i) { return 2 * i; };
-    std::cout << "\nTesting accelerator: " << alpaka::getAccName<TAcc>() << " with size: " << n << "\n"
-              << "Number of benchmark iterations: " << benchmarkIterations << "\n"
-              << "=====\n"
-              << "[[ Testing without constant iterator ]]\n";
-
-    // Use chrono library to measure time;
-    using std::chrono::duration;
-    using std::chrono::high_resolution_clock;
-
-    // REDUCE CALL:
-    // Takes the arguments: accelerator device, host device, accelerator queue, size of data, pointer-like to memory,
-    // reduce lambda.
-    // Run this once as a warm-up
-    TRed reduceResult
-        = vikunja::reduce::deviceReduce<TAcc>(devAcc, devHost, queueAcc, n, alpaka::getPtrNative(deviceMem), sum);
-
-    // Start timer
-    auto timerStart = high_resolution_clock::now();
-
-    for(int i = 0; i < benchmarkIterations; ++i)
-    {
-        // REDUCE CALL:
-        reduceResult
-            = vikunja::reduce::deviceReduce<TAcc>(devAcc, devHost, queueAcc, n, alpaka::getPtrNative(deviceMem), sum);
-    }
-
-    // End timer and count duration
-    auto timerEnd = high_resolution_clock::now();
-    duration<double, std::milli> msDouble = timerEnd - timerStart;
-
-    // check reduce result
-    auto expectedResult = n * constantIterVal;
-    std::cout << "Expected reduce result: " << expectedResult << ", real result: " << reduceResult << "\n"
-              << "Duration: " << msDouble.count() / benchmarkIterations << "ms\n";
-
-    // TRANSFORM_REDUCE CALL:
-    // Takes the arguments: accelerator device, host device, accelerator queue, size of data, pointer-like to memory,
-    // transform lambda, reduce lambda.
-    // Run this once as a warm-up
-    TRed transformReduceResult = vikunja::reduce::deviceTransformReduce<TAcc>(
-        devAcc,
-        devHost,
-        queueAcc,
-        n,
-        alpaka::getPtrNative(deviceMem),
-        doubleNum,
-        sum);
-
-    // Start timer
-    timerStart = high_resolution_clock::now();
-
-    for(int i = 0; i < benchmarkIterations; ++i)
-    {
-        // TRANSFORM_REDUCE CALL:
-        transformReduceResult = vikunja::reduce::deviceTransformReduce<TAcc>(
-            devAcc,
-            devHost,
-            queueAcc,
-            n,
-            alpaka::getPtrNative(deviceMem),
-            doubleNum,
-            sum);
-    }
-
-    // End timer and count duration
-    timerEnd = high_resolution_clock::now();
-    msDouble = timerEnd - timerStart;
-
-    // check transform result
-    auto expectedTransformReduce = expectedResult * 2;
-    std::cout << "Expected transform_reduce result: " << expectedTransformReduce
-              << ", real result: " << transformReduceResult << "\n"
-              << "Duration: " << msDouble.count() / benchmarkIterations << "ms\n"
-              << "-----\n"
-              << "[[ Testing constant iterator with value: " << constantIterVal << " ]]\n";
-
-    // Create the constant iterator
-    vikunja::mem::iterator::ConstantIterator<TRed> constantIter(constantIterVal);
-
-    // REDUCE CALL:
-    // Run this once as a warm-up
-    reduceResult = vikunja::reduce::deviceReduce<TAcc>(devAcc, devHost, queueAcc, n, constantIter, sum);
-
-    // Start timer
-    timerStart = high_resolution_clock::now();
-
-    for(int i = 0; i < benchmarkIterations; ++i)
-    {
-        // REDUCE CALL:
-        reduceResult = vikunja::reduce::deviceReduce<TAcc>(devAcc, devHost, queueAcc, n, constantIter, sum);
-    }
-
-    // End timer and count duration
-    timerEnd = high_resolution_clock::now();
-    msDouble = timerEnd - timerStart;
-
-    // check reduce result
-    expectedResult = n * constantIterVal;
-    std::cout << "Expected reduce result: " << expectedResult << ", real result: " << reduceResult << "\n"
-              << "Duration: " << msDouble.count() / benchmarkIterations << "ms\n";
-
-    // TRANSFORM_REDUCE CALL:
-    // Run this once as a warm-up
-    transformReduceResult
-        = vikunja::reduce::deviceTransformReduce<TAcc>(devAcc, devHost, queueAcc, n, constantIter, doubleNum, sum);
-
-    // Start timer
-    timerStart = high_resolution_clock::now();
-
-    for(int i = 0; i < benchmarkIterations; ++i)
-    {
-        // TRANSFORM_REDUCE CALL:
-        transformReduceResult
-            = vikunja::reduce::deviceTransformReduce<TAcc>(devAcc, devHost, queueAcc, n, constantIter, doubleNum, sum);
-    }
-
-    // End timer and count duration
-    timerEnd = high_resolution_clock::now();
-    msDouble = timerEnd - timerStart;
-
-    // check transform result
-    expectedTransformReduce = expectedResult * 2;
-    std::cout << "Expected transform_reduce result: " << expectedTransformReduce
-              << ", real result: " << transformReduceResult << "\n"
-              << "Duration: " << msDouble.count() / benchmarkIterations << "ms\n\n";
-
-    return 0;
-}
diff --git a/include/vikunja/mem/iterator/ConstantIterator.hpp b/include/vikunja/mem/iterator/ConstantIterator.hpp
index 97ff6e5..cd86f20 100644
--- a/include/vikunja/mem/iterator/ConstantIterator.hpp
+++ b/include/vikunja/mem/iterator/ConstantIterator.hpp
@@ -26,146 +26,142 @@
 #    define NODISCARD
 #endif
 
-namespace vikunja
+namespace vikunja::iterator
 {
-    namespace mem
+    /**
+     * @brief A constant iterator, returning a value given initially at any index. As such it has no bounds,
+     * other than the bounds of the index type used.
+     * @tparam DataType The type of the data
+     * @tparam IdxType The type of the index
+     */
+    template<typename DataType, typename IdxType = int64_t>
+    class ConstantIterator
     {
-        namespace iterator
-        {
-            /**
-             * @brief A constant iterator, returning a value given initially at any index. As such it has no bounds,
-             * other than the bounds of the index type used.
-             * @tparam DataType The type of the data
-             * @tparam IdxType The type of the index
-             */
-            template<typename DataType, typename IdxType = int64_t>
-            class ConstantIterator
-            {
-            public:
-                // Need all 5 of these types for iterator_traits
-                using difference_type = IdxType;
-                using value_type = DataType;
-                using pointer = DataType*;
-                using reference = DataType&;
-                using iterator_category = std::random_access_iterator_tag;
-
-                /**
-                 * @brief Constructor for the ConstantIterator
-                 * @param value The value to initialize the iterator with
-                 * @param idx The index for the iterator, default 0
-                 */
-                ALPAKA_FN_HOST_ACC constexpr ConstantIterator(const DataType& value, const IdxType& idx = {})
-                    : v(value)
-                    , index(idx)
-                {
-                }
-
-                /**
-                 * @brief Dereference operator to receive the stored value
-                 */
-                ALPAKA_FN_HOST_ACC NODISCARD constexpr ALPAKA_FN_INLINE const DataType& operator*() const
-                {
-                    return v;
-                }
-
-                /**
-                 * @brief Index operator to get stored value at some given offset from this iterator
-                 */
-                ALPAKA_FN_HOST_ACC NODISCARD constexpr ALPAKA_FN_INLINE const DataType& operator[](int) const
-                {
-                    return v;
-                }
+    public:
+        // Need all 5 of these types for iterator_traits
+        using difference_type = IdxType;
+        using value_type = DataType;
+        using pointer = DataType*;
+        using reference = DataType&;
+        using iterator_category = std::random_access_iterator_tag;
+
+        /**
+         * @brief Constructor for the ConstantIterator
+         * @param value The value to initialize the iterator with
+         * @param idx The index for the iterator, default 0
+         */
+        ALPAKA_FN_HOST_ACC constexpr ConstantIterator(const DataType& value, const IdxType& idx = {})
+            : v(value)
+            , index(idx)
+        {
+        }
+
+        /**
+         * @brief Dereference operator to receive the stored value
+         */
+        ALPAKA_FN_HOST_ACC NODISCARD constexpr ALPAKA_FN_INLINE const DataType& operator*() const
+        {
+            return v;
+        }
+
+        /**
+         * @brief Index operator to get stored value at some given offset from this iterator
+         */
+        ALPAKA_FN_HOST_ACC NODISCARD constexpr ALPAKA_FN_INLINE const DataType& operator[](int) const
+        {
+            return v;
+        }
 
 #pragma region arithmeticoperators
-                /**
-                 * @brief Postfix increment operator
-                 * @note Use prefix increment operator instead if possible to avoid copies
-                 */
-                ALPAKA_FN_HOST_ACC constexpr ALPAKA_FN_INLINE ConstantIterator operator++()
-                {
-                    ConstantIterator cpy = *this;
-                    ++index;
-                    return cpy;
-                }
-
-                /**
-                 * @brief Prefix increment operator
-                 */
-                ALPAKA_FN_HOST_ACC constexpr ALPAKA_FN_INLINE ConstantIterator& operator++(int)
-                {
-                    ++index;
-                    return *this;
-                }
-
-                /**
-                 * @brief Postfix decrement operator
-                 * @note Use prefix decrement operator instead if possible to avoid copies
-                 */
-                ALPAKA_FN_HOST_ACC constexpr ALPAKA_FN_INLINE ConstantIterator operator--()
-                {
-                    ConstantIterator cpy = *this;
-                    --index;
-                    return cpy;
-                }
-
-                /**
-                 * @brief Prefix decrement operator
-                 */
-                ALPAKA_FN_HOST_ACC constexpr ALPAKA_FN_INLINE ConstantIterator& operator--(int)
-                {
-                    --index;
-                    return *this;
-                }
-
-                /**
-                 * @brief Add an index to this iterator
-                 */
-                ALPAKA_FN_HOST_ACC NODISCARD constexpr friend ALPAKA_FN_INLINE ConstantIterator
-                operator+(ConstantIterator it, IdxType idx)
-                {
-                    return it += idx;
-                }
-
-                /**
-                 * @brief Subtract an index from this iterator
-                 */
-                ALPAKA_FN_HOST_ACC NODISCARD constexpr friend ALPAKA_FN_INLINE ConstantIterator
-                operator-(ConstantIterator it, const IdxType idx)
-                {
-                    return it -= idx;
-                }
-
-                /**
-                 * @brief Subtract a second constant iterator of the same value from this one
-                 */
-                ALPAKA_FN_HOST_ACC NODISCARD constexpr friend ALPAKA_FN_INLINE IdxType
-                operator-(ConstantIterator it, const ConstantIterator& other)
-                {
-                    assert(it.v == other.v && "Can't subtract constant iterators of different values!");
-                    return it.index - other.index;
-                }
-
-                /**
-                 * @brief Add an index to this iterator
-                 */
-                ALPAKA_FN_HOST_ACC constexpr friend ALPAKA_FN_INLINE ConstantIterator& operator+=(
-                    ConstantIterator& it,
-                    const IdxType idx)
-                {
-                    it.index += idx;
-                    return it;
-                }
-
-                /**
-                 * @brief Subtract an index from this iterator
-                 */
-                ALPAKA_FN_HOST_ACC constexpr friend ALPAKA_FN_INLINE ConstantIterator& operator-=(
-                    ConstantIterator& it,
-                    const IdxType idx)
-                {
-                    it.index -= idx;
-                    return it;
-                }
+        /**
+         * @brief Postfix increment operator
+         * @note Use prefix increment operator instead if possible to avoid copies
+         */
+        ALPAKA_FN_HOST_ACC constexpr ALPAKA_FN_INLINE ConstantIterator operator++()
+        {
+            ConstantIterator cpy = *this;
+            ++index;
+            return cpy;
+        }
+
+        /**
+         * @brief Prefix increment operator
+         */
+        ALPAKA_FN_HOST_ACC constexpr ALPAKA_FN_INLINE ConstantIterator& operator++(int)
+        {
+            ++index;
+            return *this;
+        }
+
+        /**
+         * @brief Postfix decrement operator
+         * @note Use prefix decrement operator instead if possible to avoid copies
+         */
+        ALPAKA_FN_HOST_ACC constexpr ALPAKA_FN_INLINE ConstantIterator operator--()
+        {
+            ConstantIterator cpy = *this;
+            --index;
+            return cpy;
+        }
+
+        /**
+         * @brief Prefix decrement operator
+         */
+        ALPAKA_FN_HOST_ACC constexpr ALPAKA_FN_INLINE ConstantIterator& operator--(int)
+        {
+            --index;
+            return *this;
+        }
+
+        /**
+         * @brief Add an index to this iterator
+         */
+        ALPAKA_FN_HOST_ACC NODISCARD constexpr friend ALPAKA_FN_INLINE ConstantIterator
+        operator+(ConstantIterator it, IdxType idx)
+        {
+            return it += idx;
+        }
+
+        /**
+         * @brief Subtract an index from this iterator
+         */
+        ALPAKA_FN_HOST_ACC NODISCARD constexpr friend ALPAKA_FN_INLINE ConstantIterator
+        operator-(ConstantIterator it, const IdxType idx)
+        {
+            return it -= idx;
+        }
+
+        /**
+         * @brief Subtract a second constant iterator of the same value from this one
+         */
+        ALPAKA_FN_HOST_ACC NODISCARD constexpr friend ALPAKA_FN_INLINE IdxType
+        operator-(ConstantIterator it, const ConstantIterator& other)
+        {
+            assert(it.v == other.v && "Can't subtract constant iterators of different values!");
+            return it.index - other.index;
+        }
+
+        /**
+         * @brief Add an index to this iterator
+         */
+        ALPAKA_FN_HOST_ACC constexpr friend ALPAKA_FN_INLINE ConstantIterator& operator+=(
+            ConstantIterator& it,
+            const IdxType idx)
+        {
+            it.index += idx;
+            return it;
+        }
+
+        /**
+         * @brief Subtract an index from this iterator
+         */
+        ALPAKA_FN_HOST_ACC constexpr friend ALPAKA_FN_INLINE ConstantIterator& operator-=(
+            ConstantIterator& it,
+            const IdxType idx)
+        {
+            it.index -= idx;
+            return it;
+        }
 
 #pragma endregion arithmeticoperators
 
@@ -173,65 +169,63 @@ namespace vikunja
 
 #ifdef USESPACESHIP
 
-                ALPAKA_FN_HOST_ACC NODISCARD constexpr ALPAKA_FN_INLINE auto operator<=>(
-                    const ConstantIterator& other) const noexcept = default;
+        ALPAKA_FN_HOST_ACC NODISCARD constexpr ALPAKA_FN_INLINE auto operator<=>(
+            const ConstantIterator& other) const noexcept = default;
 
 #else
 
-                ALPAKA_FN_HOST_ACC NODISCARD constexpr friend ALPAKA_FN_INLINE bool operator==(
-                    const ConstantIterator& it,
-                    const ConstantIterator& other) noexcept
-                {
-                    return it.v == other.v && it.index == other.index;
-                }
-
-                ALPAKA_FN_HOST_ACC NODISCARD constexpr friend ALPAKA_FN_INLINE bool operator!=(
-                    const ConstantIterator& it,
-                    const ConstantIterator& other) noexcept
-                {
-                    return !operator==(it, other);
-                }
-
-                ALPAKA_FN_HOST_ACC NODISCARD constexpr friend ALPAKA_FN_INLINE bool operator<(
-                    const ConstantIterator& it,
-                    const ConstantIterator& other) noexcept
-                {
-                    if(it.v < other.v)
-                        return true;
-                    if(it.v > other.v)
-                        return false;
-                    return it.index < other.index;
-                }
-
-                ALPAKA_FN_HOST_ACC NODISCARD constexpr friend ALPAKA_FN_INLINE bool operator>(
-                    const ConstantIterator& it,
-                    const ConstantIterator& other) noexcept
-                {
-                    return operator<(other, it);
-                }
-
-                ALPAKA_FN_HOST_ACC NODISCARD constexpr friend ALPAKA_FN_INLINE bool operator<=(
-                    const ConstantIterator& it,
-                    const ConstantIterator& other) noexcept
-                {
-                    return operator<(it, other) || operator==(it, other);
-                }
-
-                ALPAKA_FN_HOST_ACC NODISCARD constexpr friend ALPAKA_FN_INLINE bool operator>=(
-                    const ConstantIterator& it,
-                    const ConstantIterator& other) noexcept
-                {
-                    return operator<=(other, it);
-                }
+        ALPAKA_FN_HOST_ACC NODISCARD constexpr friend ALPAKA_FN_INLINE bool operator==(
+            const ConstantIterator& it,
+            const ConstantIterator& other) noexcept
+        {
+            return it.v == other.v && it.index == other.index;
+        }
+
+        ALPAKA_FN_HOST_ACC NODISCARD constexpr friend ALPAKA_FN_INLINE bool operator!=(
+            const ConstantIterator& it,
+            const ConstantIterator& other) noexcept
+        {
+            return !operator==(it, other);
+        }
+
+        ALPAKA_FN_HOST_ACC NODISCARD constexpr friend ALPAKA_FN_INLINE bool operator<(
+            const ConstantIterator& it,
+            const ConstantIterator& other) noexcept
+        {
+            if(it.v < other.v)
+                return true;
+            if(it.v > other.v)
+                return false;
+            return it.index < other.index;
+        }
+
+        ALPAKA_FN_HOST_ACC NODISCARD constexpr friend ALPAKA_FN_INLINE bool operator>(
+            const ConstantIterator& it,
+            const ConstantIterator& other) noexcept
+        {
+            return operator<(other, it);
+        }
+
+        ALPAKA_FN_HOST_ACC NODISCARD constexpr friend ALPAKA_FN_INLINE bool operator<=(
+            const ConstantIterator& it,
+            const ConstantIterator& other) noexcept
+        {
+            return operator<(it, other) || operator==(it, other);
+        }
+
+        ALPAKA_FN_HOST_ACC NODISCARD constexpr friend ALPAKA_FN_INLINE bool operator>=(
+            const ConstantIterator& it,
+            const ConstantIterator& other) noexcept
+        {
+            return operator<=(other, it);
+        }
 #endif
 
 #pragma endregion comparisonoperators
 
-            private:
-                DataType v;
-                IdxType index;
-            };
+    private:
+        DataType v;
+        IdxType index;
+    };
 
-        } // namespace iterator
-    } // namespace mem
-} // namespace vikunja
+} // namespace vikunja::iterator
diff --git a/test/benchmarks/reduce/bench_vikunja_constiter_reduce.cpp b/test/benchmarks/reduce/bench_vikunja_constiter_reduce.cpp
index e4d7bd8..b8c42e2 100644
--- a/test/benchmarks/reduce/bench_vikunja_constiter_reduce.cpp
+++ b/test/benchmarks/reduce/bench_vikunja_constiter_reduce.cpp
@@ -34,7 +34,7 @@ inline void reduce_benchmark(TIdx size)
         >;
     using Vec = alpaka::Vec<typename Setup::Dim, typename Setup::Idx>;
 
-    using ConstantIterator = vikunja::mem::iterator::ConstantIterator<TData>;
+    using ConstantIterator = vikunja::iterator::ConstantIterator<TData>;
 
     INFO((vikunja::test::print_acc_info<typename Setup::Dim>(size)));
 
diff --git a/test/integ/transform/src/ConstantIterator.cpp b/test/integ/transform/src/ConstantIterator.cpp
index 7553725..babe5a9 100644
--- a/test/integ/transform/src/ConstantIterator.cpp
+++ b/test/integ/transform/src/ConstantIterator.cpp
@@ -64,7 +64,7 @@ TEST_CASE("ConstantIteratorTest", "[transform][iterator][lambda]")
     auto deviceOutMem(alpaka::allocBuf<TTrans, Setup::Idx>(setup.devAcc, extent));
     auto hostOutMem(alpaka::allocBuf<TTrans, Setup::Idx>(setup.devHost, extent));
 
-    vikunja::mem::iterator::ConstantIterator c_begin(constantIterVal);
+    vikunja::iterator::ConstantIterator c_begin(constantIterVal);
 
     vikunja::transform::deviceTransform<Setup::Acc>(
         setup.devAcc,
diff --git a/test/unit/iterator/src/ConstantIterator.cpp b/test/unit/iterator/src/ConstantIterator.cpp
index 5ac069f..9382ddb 100644
--- a/test/unit/iterator/src/ConstantIterator.cpp
+++ b/test/unit/iterator/src/ConstantIterator.cpp
@@ -13,7 +13,7 @@
 
 #include <catch2/catch.hpp>
 
-using vikunja::mem::iterator::ConstantIterator;
+using vikunja::iterator::ConstantIterator;
 
 TEMPLATE_TEST_CASE("Test Constant Iterator", "", int, double, uint64_t)
 {