AMReX_RealVect.H

Go to the documentation of this file.

#ifndef AMREX_REALVECT_H_
#define AMREX_REALVECT_H_
#include <AMReX_Config.H>
 
#include <AMReX_Box.H>
#include <AMReX_REAL.H>
#include <AMReX_SPACE.H>
#include <AMReX_IntVect.H>
#include <AMReX_Utility.H>
#include <AMReX_Math.H>
 
#include <cmath>
#include <cstddef>
#include <cstdlib>
#include <cstring>
#include <iosfwd>
#include <vector>
 
namespace amrex
{
 
template <int dim>
class RealVectND {
    public:
 
 
    constexpr RealVectND () noexcept {}  // NOLINT
 
    explicit RealVectND (const std::vector<Real>& vr) noexcept {
        BL_ASSERT(vr.size() == dim);
        for (int i = 0; i < dim; ++i) {
            vect[i] = vr[i];
        }
    }
 
 
    template <class... Args,
        std::enable_if_t<
            (sizeof...(Args) + 2 == dim) &&
            IsConvertible_v<Real, Args...>,
        int> = 0>
    AMREX_GPU_HOST_DEVICE
    constexpr RealVectND (Real i, Real j, Args... ks) noexcept
        : vect{i, j, static_cast<Real>(ks)...} {}
 
    AMREX_GPU_HOST_DEVICE
    explicit constexpr RealVectND (Real s) noexcept {
        for (int i = 0; i < dim; ++i) {
            vect[i] = s;
        }
    }
 
 
    AMREX_GPU_HOST_DEVICE
    explicit RealVectND (const Real* a) noexcept {
        for (int i = 0; i < dim; ++i) {
            vect[i] = a[i];
        }
    }
 
 
    AMREX_GPU_HOST_DEVICE constexpr
    RealVectND (const IntVectND<dim>& iv) noexcept {
        for (int i = 0; i < dim; ++i) {
            vect[i] = static_cast<Real>(iv[i]);
        }
    }
 
 
   AMREX_GPU_HOST_DEVICE constexpr
   RealVectND (const GpuArray<Real, dim> ga) noexcept {
       for (int i = 0; i < dim; ++i) {
           vect[i] = ga[i];
       }
   }
 
 
    Real& operator[] (int i) && = delete;
 
 
    AMREX_GPU_HOST_DEVICE inline constexpr
    Real& operator[] (int i) & noexcept {
        AMREX_ASSERT(i >= 0 && i < dim);
        return vect[i];
    }
 
 
    AMREX_GPU_HOST_DEVICE inline constexpr
    const Real& operator[] (int i) const& noexcept {
        AMREX_ASSERT(i >= 0 && i < dim);
        return vect[i];
    }
 
    template<std::size_t i>
    [[nodiscard]] AMREX_GPU_HOST_DEVICE AMREX_FORCE_INLINE constexpr
    Real& get () noexcept {static_assert(0<=i && i<dim); return vect[i];}
 
    template<std::size_t i>
    [[nodiscard]] AMREX_GPU_HOST_DEVICE AMREX_FORCE_INLINE constexpr
    const Real& get () const noexcept {static_assert(0<=i && i<dim); return vect[i];}
 
    [[nodiscard]] AMREX_GPU_HOST_DEVICE inline constexpr
    GpuArray<Real, dim> to_array () const noexcept {
        GpuArray<Real, dim> retval{};
        for (int i = 0; i < dim; ++i) {
            retval[i] = vect[i];
        }
        return retval;
    }
 
 
 
    [[nodiscard]] AMREX_GPU_HOST_DEVICE inline constexpr
    Real* begin () noexcept { return vect; }
 
 
    [[nodiscard]] AMREX_GPU_HOST_DEVICE inline constexpr
    const Real* begin () const noexcept { return vect; }
 
 
    [[nodiscard]] AMREX_GPU_HOST_DEVICE inline constexpr
    Real* end () noexcept { return vect + dim; }
 
 
    [[nodiscard]] AMREX_GPU_HOST_DEVICE inline constexpr
    const Real* end () const noexcept {
        return vect + dim;
    }
 
 
 
    AMREX_GPU_HOST_DEVICE constexpr
    bool operator== (const RealVectND& p) const noexcept {
        bool retval = vect[0] == p[0];
        for (int i=1; i<dim; ++i) {
            retval = retval && vect[i] == p[i];
        }
        return retval;
    }
 
 
    AMREX_GPU_HOST_DEVICE constexpr
    bool operator!= (const RealVectND& p) const noexcept {
        bool retval = vect[0] != p[0];
        for (int i=1; i<dim; ++i) {
            retval = retval || vect[i] != p[i];
        }
        return retval;
    }
 
 
    AMREX_GPU_HOST_DEVICE inline bool operator< (const RealVectND& p) const noexcept;
 
 
    AMREX_GPU_HOST_DEVICE inline bool operator<= (const RealVectND& p) const noexcept;
 
 
    AMREX_GPU_HOST_DEVICE inline bool operator> (const RealVectND& p) const noexcept;
 
 
    AMREX_GPU_HOST_DEVICE inline bool operator>= (const RealVectND& p) const noexcept;
 
 
 
    AMREX_GPU_HOST_DEVICE inline RealVectND& operator+= (Real s) noexcept;
 
 
    AMREX_GPU_HOST_DEVICE inline RealVectND operator+ (Real s) const noexcept;
 
 
    AMREX_GPU_HOST_DEVICE inline RealVectND& operator+= (const RealVectND& p) noexcept;
 
 
    AMREX_GPU_HOST_DEVICE inline RealVectND& operator-= (Real s) noexcept;
 
 
    AMREX_GPU_HOST_DEVICE inline RealVectND& operator-= (const RealVectND& p) noexcept;
 
 
    AMREX_GPU_HOST_DEVICE inline RealVectND operator- (Real s) const noexcept;
 
 
    AMREX_GPU_HOST_DEVICE inline RealVectND& operator*= (Real s) noexcept;
 
 
    [[nodiscard]] AMREX_GPU_HOST_DEVICE inline Real dotProduct (
        const RealVectND& a_rhs) const noexcept;
 
 
    template <int N=dim, std::enable_if_t<( N==3 ), int> = 0>
    [[nodiscard]] AMREX_GPU_HOST_DEVICE inline RealVectND crossProduct (
        const RealVectND& a_rhs) const noexcept;
 
 
    AMREX_GPU_HOST_DEVICE inline RealVectND& operator*= (const RealVectND& p) noexcept;
 
 
    AMREX_GPU_HOST_DEVICE inline RealVectND operator* (Real s) const noexcept;
 
 
    AMREX_GPU_HOST_DEVICE inline RealVectND& operator/= (Real s) noexcept;
 
 
    AMREX_GPU_HOST_DEVICE inline RealVectND& operator/= (const RealVectND& p) noexcept;
 
 
    AMREX_GPU_HOST_DEVICE inline RealVectND operator/ (Real s) const noexcept;
 
 
    AMREX_GPU_HOST_DEVICE inline RealVectND& scale (Real s) noexcept;
 
 
    [[nodiscard]] AMREX_GPU_HOST_DEVICE inline IntVectND<dim> floor () const noexcept;
 
 
    [[nodiscard]] AMREX_GPU_HOST_DEVICE inline IntVectND<dim> ceil () const noexcept;
 
 
    [[nodiscard]] AMREX_GPU_HOST_DEVICE inline IntVectND<dim> round () const noexcept;
 
 
 
    AMREX_GPU_HOST_DEVICE inline RealVectND& min (const RealVectND& p) noexcept;
 
 
    AMREX_GPU_HOST_DEVICE inline RealVectND& max (const RealVectND& p) noexcept;
 
 
 
    AMREX_GPU_HOST_DEVICE inline RealVectND operator+ () const noexcept;
 
 
    AMREX_GPU_HOST_DEVICE inline RealVectND operator- () const noexcept;
 
 
    [[nodiscard]] AMREX_GPU_HOST_DEVICE inline Real sum () const noexcept;
 
 
    [[nodiscard]] AMREX_GPU_HOST_DEVICE inline Real vectorLength () const noexcept;
 
 
    [[nodiscard]] AMREX_GPU_HOST_DEVICE inline Real radSquared () const noexcept;
 
 
    [[nodiscard]] AMREX_GPU_HOST_DEVICE inline Real product () const noexcept;
 
 
    [[nodiscard]] AMREX_GPU_HOST_DEVICE inline int minDir (const bool& a_doAbs) const noexcept;
 
 
    [[nodiscard]] AMREX_GPU_HOST_DEVICE inline int maxDir (const bool& a_doAbs) const noexcept;
 
 
 
    [[nodiscard]] AMREX_GPU_HOST_DEVICE const Real* dataPtr () const noexcept { return vect; }
 
 
    AMREX_GPU_HOST_DEVICE
    Real* dataPtr () noexcept { return vect; }
 
 
 
    static const RealVectND Zero;
 
 
    static const RealVectND Unit;
 
 
 
    AMREX_GPU_HOST_DEVICE AMREX_FORCE_INLINE static constexpr RealVectND TheZeroVector () noexcept {
        return RealVectND{0.};
    }
 
    AMREX_GPU_HOST_DEVICE AMREX_FORCE_INLINE static constexpr RealVectND TheUnitVector () noexcept {
        return RealVectND{1.};
    }
 
    AMREX_GPU_HOST_DEVICE AMREX_FORCE_INLINE
    static constexpr std::size_t size () noexcept {
        return static_cast<std::size_t>(dim);
    }
 
    AMREX_GPU_HOST_DEVICE AMREX_FORCE_INLINE
    static constexpr int isize () noexcept {
        return dim;
    }
 
    using value_type = Real;
 
    template<int new_dim>
    [[nodiscard]] AMREX_GPU_HOST_DEVICE AMREX_FORCE_INLINE
    RealVectND<new_dim> shrink () const noexcept {
        static_assert(new_dim <= dim);
        return RealVectND<new_dim>(this->begin());
    }
 
    template<int new_dim>
    [[nodiscard]] AMREX_GPU_HOST_DEVICE AMREX_FORCE_INLINE
    RealVectND<new_dim> expand (Real fill_extra=0.) const noexcept {
        static_assert(new_dim >= dim);
        RealVectND<new_dim> retval(fill_extra);
        for (int i=0; i<dim; ++i) {
            retval[i] = vect[i];
        }
        return retval;
    }
 
    template<int new_dim>
    [[nodiscard]] AMREX_GPU_HOST_DEVICE AMREX_FORCE_INLINE
    RealVectND<new_dim> resize (Real fill_extra=0.) const noexcept {
        if constexpr (new_dim > dim) {
            return expand<new_dim>(fill_extra);
        } else {
            return shrink<new_dim>();
        }
    }
 
protected:
    Real vect[dim] = {};
};
 
// Template deduction guide for RealVectND
template <class...Args,
    std::enable_if_t<
        IsConvertible_v<Real, Args...>,
    int> = 0>
AMREX_GPU_HOST_DEVICE // __device__ for HIP
RealVectND(Real, Real, Args...) -> RealVectND<sizeof...(Args)+2>;
 
// Template deduction guide for RealVectND
template<int dim>
AMREX_GPU_HOST_DEVICE // __device__ for HIP
RealVectND(const IntVectND<dim>&) -> RealVectND<dim>;
 
// Template deduction guide for RealVectND
template<int dim>
AMREX_GPU_HOST_DEVICE // __device__ for HIP
RealVectND(const GpuArray<Real, dim>&) -> RealVectND<dim>;
 
template <int dim>
inline constexpr const RealVectND<dim> RealVectND<dim>::Zero{Real(0)};
 
template <int dim>
inline constexpr const RealVectND<dim> RealVectND<dim>::Unit{Real(1)};
 
using RealVect = RealVectND<AMREX_SPACEDIM>;
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim>&
RealVectND<dim>::operator-= (Real s) noexcept {
    for (int i=0; i<dim; ++i) {
        vect[i] -= s;
    }
    return *this;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim>&
RealVectND<dim>::operator*= (Real s) noexcept {
    for (int i=0; i<dim; ++i) {
        vect[i] *= s;
    }
    return *this;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim>&
RealVectND<dim>::operator-= (const RealVectND<dim>& p) noexcept {
    for (int i=0; i<dim; ++i) {
        vect[i] -= p[i];
    }
    return *this;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim>
RealVectND<dim>::operator+ () const noexcept {
    return RealVectND<dim>(*this);
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim>
RealVectND<dim>::operator- () const noexcept {
    RealVectND<dim> retval;
    for (int i=0; i<dim; ++i) {
        retval[i] = -vect[i];
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim>&
RealVectND<dim>::scale (Real s) noexcept {
    for (int i=0; i<dim; ++i) {
        vect[i] *= s;
    }
    return *this;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
IntVectND<dim>
RealVectND<dim>::floor () const noexcept {
    IntVectND<dim> retval;
    for (int i=0; i<dim; ++i) {
        retval[i] = static_cast<int>(std::floor(vect[i]));
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
IntVectND<dim>
RealVectND<dim>::ceil () const noexcept {
    IntVectND<dim> retval;
    for (int i=0; i<dim; ++i) {
        retval[i] = static_cast<int>(std::ceil(vect[i]));
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
IntVectND<dim>
RealVectND<dim>::round () const noexcept {
    IntVectND<dim> retval;
    for (int i=0; i<dim; ++i) {
        retval[i] = static_cast<int>(std::round(vect[i]));
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
Real
RealVectND<dim>::sum () const noexcept {
    Real retval = vect[0];
    for (int i=1; i<dim; ++i) {
        retval += vect[i];
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
Real
RealVectND<dim>::vectorLength () const noexcept {
    Real len = this->radSquared();
    len = std::sqrt(len);
    return len;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
Real
RealVectND<dim>::radSquared () const noexcept {
    Real retval = vect[0] * vect[0];
    for (int i=1; i<dim; ++i) {
        retval += vect[i] * vect[i];
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
Real
RealVectND<dim>::product () const noexcept {
    Real retval = vect[0];
    for (int i=1; i<dim; ++i) {
        retval *= vect[i];
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
bool
RealVectND<dim>::operator< (const RealVectND<dim>& p) const noexcept {
    bool retval = vect[0] < p[0];
    for (int i=1; i<dim; ++i) {
        retval = retval && vect[i] < p[i];
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
bool
RealVectND<dim>::operator<= (const RealVectND<dim>& p) const noexcept {
    bool retval = vect[0] <= p[0];
    for (int i=1; i<dim; ++i) {
        retval = retval && vect[i] <= p[i];
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
bool
RealVectND<dim>::operator> (const RealVectND<dim>& p) const noexcept {
    bool retval = vect[0] > p[0];
    for (int i=1; i<dim; ++i) {
        retval = retval && vect[i] > p[i];
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
bool
RealVectND<dim>::operator>= (const RealVectND<dim>& p) const noexcept {
    bool retval = vect[0] >= p[0];
    for (int i=1; i<dim; ++i) {
        retval = retval && vect[i] >= p[i];
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim>&
RealVectND<dim>::min (const RealVectND<dim>& p) noexcept {
    for (int i=0; i<dim; ++i) {
        vect[i] = amrex::min(vect[i], p.vect[i]);
    }
    return *this;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim>&
RealVectND<dim>::max (const RealVectND<dim>& p) noexcept {
    for (int i=0; i<dim; ++i) {
        vect[i] = amrex::max(vect[i], p.vect[i]);
    }
    return *this;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim> min (const RealVectND<dim>& p1, const RealVectND<dim>& p2) noexcept {
    RealVectND<dim> p(p1);
    return p.min(p2);
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim> max (const RealVectND<dim>& p1, const RealVectND<dim>& p2) noexcept {
    RealVectND<dim> p(p1);
    return p.max(p2);
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
Real
RealVectND<dim>::dotProduct (const RealVectND<dim>& a_rhs) const noexcept {
    Real retval  = vect[0] * a_rhs.vect[0];
    for (int i=1; i<dim; ++i) {
        retval += vect[i] * a_rhs.vect[i];
    }
    return retval;
}
 
template <int dim>
template <int N, std::enable_if_t<( N==3 ), int>>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim>
RealVectND<dim>::crossProduct (const RealVectND<dim>& a_rhs) const noexcept {
    RealVectND<dim> tmp(vect[1] * a_rhs[2] - vect[2] * a_rhs[1],
                        vect[2] * a_rhs[0] - vect[0] * a_rhs[2],
                        vect[0] * a_rhs[1] - vect[1] * a_rhs[0]);
    return tmp;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim>&
RealVectND<dim>::operator+= (Real s) noexcept {
    for (int i=0; i<dim; ++i) {
        vect[i] += s;
    }
    return *this;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim>&
RealVectND<dim>::operator+= (const RealVectND<dim>& p) noexcept {
    for (int i=0; i<dim; ++i) {
        vect[i] += p[i];
    }
    return *this;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim>&
RealVectND<dim>::operator*= (const RealVectND<dim>& p) noexcept {
    for (int i=0; i<dim; ++i) {
        vect[i] *= p[i];
    }
    return *this;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim>
RealVectND<dim>::operator* (Real s) const noexcept {
    RealVectND<dim> retval;
    for (int i=0; i<dim; ++i) {
        retval[i] = vect[i] * s;
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim>
RealVectND<dim>::operator- (Real s) const noexcept {
    RealVectND<dim> retval;
    for (int i=0; i<dim; ++i) {
        retval[i] = vect[i] - s;
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim>
RealVectND<dim>::operator+ (Real s) const noexcept {
    RealVectND<dim> retval;
    for (int i=0; i<dim; ++i) {
        retval[i] = vect[i] + s;
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim>&
RealVectND<dim>::operator/= (Real s) noexcept {
    for (int i=0; i<dim; ++i) {
        vect[i] /= s;
    }
    return *this;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim>&
RealVectND<dim>::operator/= (const RealVectND<dim>& p) noexcept {
    for (int i=0; i<dim; ++i) {
        vect[i] /= p[i];
    }
    return *this;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim>
RealVectND<dim>::operator/ (Real s) const noexcept {
    RealVectND<dim> retval;
    for (int i=0; i<dim; ++i) {
        retval[i] = vect[i] / s;
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
int
RealVectND<dim>::minDir (const bool& a_doAbs) const noexcept {
    int mDir = 0;
    for (int idir = 0; idir < dim; idir++) {
        if (a_doAbs) {
            if (std::abs(vect[idir]) < std::abs(vect[mDir])) {
                mDir = idir;
            }
        } else {
            if (vect[idir] < vect[mDir]) {
                mDir = idir;
            }
        }
    }
    return mDir;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
int
RealVectND<dim>::maxDir (const bool& a_doAbs) const noexcept {
    int mDir = 0;
    for (int idir = 0; idir < dim; idir++) {
        if (a_doAbs) {
            if (std::abs(vect[idir]) > std::abs(vect[mDir])) {
                mDir = idir;
            }
        } else {
            if (vect[idir] > vect[mDir]) {
                mDir = idir;
            }
        }
    }
    return mDir;
}
 
template <int dim=AMREX_SPACEDIM>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim> BASISREALV (int dir) noexcept {
    AMREX_ASSERT(dir >= 0 && dir < dim);
    RealVectND<dim> tmp{0.};
    tmp[dir] = 1;
    return tmp;
}
 
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim> operator/ (Real s, const RealVectND<dim>& p) noexcept {
    RealVectND<dim> retval;
    for (int i=0; i<dim; ++i) {
        retval[i] = s / p[i];
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim> operator+ (Real s, const RealVectND<dim>& p) noexcept {
    RealVectND<dim> retval;
    for (int i=0; i<dim; ++i) {
        retval[i] = p[i] + s;
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim> operator- (Real s, const RealVectND<dim>& p) noexcept {
    RealVectND<dim> retval;
    for (int i=0; i<dim; ++i) {
        retval[i] = s - p[i];
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim> operator* (Real s, const RealVectND<dim>& p) noexcept {
    RealVectND<dim> retval;
    for (int i=0; i<dim; ++i) {
        retval[i] = s * p[i];
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim> operator/ (const RealVectND<dim>& s, const RealVectND<dim>& p) noexcept {
    RealVectND<dim> retval;
    for (int i=0; i<dim; ++i) {
        retval[i] = s[i] / p[i];
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim> operator+ (const RealVectND<dim>& s, const RealVectND<dim>& p) noexcept {
    RealVectND<dim> retval;
    for (int i=0; i<dim; ++i) {
        retval[i] = p[i] + s[i];
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim> operator- (const RealVectND<dim>& s, const RealVectND<dim>& p) noexcept {
    RealVectND<dim> retval;
    for (int i=0; i<dim; ++i) {
        retval[i] = s[i] - p[i];
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim> operator* (const RealVectND<dim>& s, const RealVectND<dim>& p) noexcept {
    RealVectND<dim> retval;
    for (int i=0; i<dim; ++i) {
        retval[i] = p[i] * s[i];
    }
    return retval;
}
 
template <int dim>
AMREX_GPU_HOST_DEVICE inline
RealVectND<dim> scale (const RealVectND<dim>& p, Real s) noexcept {
    RealVectND<dim> retval;
    for (int i=0; i<dim; ++i) {
        retval[i] = s * p[i];
    }
    return retval;
}
 
namespace detail {
    std::ostream& real_vector_write (std::ostream& os, const Real* p, int dim);
    std::istream& real_vector_read (std::istream& is, Real* p, int dim);
 
    template<int dim>
    AMREX_GPU_HOST_DEVICE AMREX_FORCE_INLINE constexpr
    void RealVectCat_imp (Real*& dst, const RealVectND<dim>& src) noexcept {
        for (int i=0; i<dim; ++i) {
            dst[i] = src[i];
        }
        dst += dim;
    }
 
    template<int dim>
    AMREX_GPU_HOST_DEVICE AMREX_FORCE_INLINE constexpr
    void RealVectSplit_imp2 (RealVectND<dim>& dst, const Real*& src) noexcept {
        for (int i=0; i<dim; ++i) {
            dst[i] = src[i];
        }
        src += dim;
    }
 
    template<class T, std::size_t...Ns>
    AMREX_GPU_HOST_DEVICE AMREX_FORCE_INLINE constexpr
    T RealVectSplit_imp (T& retval, std::index_sequence<Ns...>, const Real * src) noexcept {
        (RealVectSplit_imp2(amrex::get<Ns>(retval), src), ...);
        return retval;
    }
}
 
template<int dim>
std::ostream&
operator<< (std::ostream& os, const RealVectND<dim>& p)
{
    return detail::real_vector_write(os, p.begin(), dim);
}
 
template<int dim>
std::istream&
operator>> (std::istream& is, RealVectND<dim>& p)
{
    return detail::real_vector_read(is, p.begin(), dim);
}
 
template<int d, int...dims>
AMREX_GPU_HOST_DEVICE
AMREX_FORCE_INLINE
constexpr RealVectND<detail::get_sum<d, dims...>()>
RealVectCat (const RealVectND<d>& v, const RealVectND<dims>&...vects) noexcept {
    RealVectND<detail::get_sum<d, dims...>()> retval {0.};
    Real* dst = retval.begin();
    detail::RealVectCat_imp(dst, v);
    (detail::RealVectCat_imp(dst, vects), ...);
    return retval;
}
 
template<int d, int...dims>
AMREX_GPU_HOST_DEVICE
AMREX_FORCE_INLINE
constexpr GpuTuple<RealVectND<d>, RealVectND<dims>...>
RealVectSplit (const RealVectND<detail::get_sum<d, dims...>()>& v) noexcept {
    GpuTuple<RealVectND<d>, RealVectND<dims>...> retval;
    return detail::RealVectSplit_imp(retval,
                                    std::make_index_sequence<1 + sizeof...(dims)>(),
                                    v.begin());
}
 
template<int new_dim, int old_dim>
AMREX_GPU_HOST_DEVICE
AMREX_FORCE_INLINE
constexpr RealVectND<new_dim>
RealVectShrink (const RealVectND<old_dim>& iv) noexcept {
    return iv.template shrink<new_dim>();
}
 
template<int new_dim, int old_dim>
AMREX_GPU_HOST_DEVICE
AMREX_FORCE_INLINE
constexpr RealVectND<new_dim>
RealVectExpand (const RealVectND<old_dim>& iv, Real fill_extra=0) noexcept {
    return iv.template expand<new_dim>(fill_extra);
}
 
template<int new_dim, int old_dim>
AMREX_GPU_HOST_DEVICE
AMREX_FORCE_INLINE
constexpr RealVectND<new_dim>
RealVectResize (const RealVectND<old_dim>& iv, Real fill_extra=0) noexcept {
    return iv.template resize<new_dim>(fill_extra);
}
 
}  // namespace amrex
 
// Spcialize std::tuple_size for RealVectND. Used by structured bindings.
template<int dim>
struct std::tuple_size<amrex::RealVectND<dim>> {
    static constexpr std::size_t value = dim;
};
 
// Spcialize std::tuple_element for RealVectND. Used by structured bindings.
template<std::size_t s, int dim>
struct std::tuple_element<s, amrex::RealVectND<dim>> {
    using type = amrex::Real;
};
 
#endif