From ccf7d71eaea8b5bd106cf99406f62947cbb7b3ea Mon Sep 17 00:00:00 2001
From: Serhii Yaskovets <yaskovet@mpi-cbg.de>
Date: Thu, 22 Jun 2023 15:19:31 +0200
Subject: [PATCH] Add missing vector_dist methods for DCPSE on GPU
---
src/Vector/vector_dist_kernel.hpp | 56 +++++++++++++++++++++++++++++++
src/Vector/vector_dist_subset.hpp | 21 ++++++++++++
2 files changed, 77 insertions(+)
diff --git a/src/Vector/vector_dist_kernel.hpp b/src/Vector/vector_dist_kernel.hpp
index cc42a53e3..91276dfdb 100644
--- a/src/Vector/vector_dist_kernel.hpp
+++ b/src/Vector/vector_dist_kernel.hpp
@@ -167,6 +167,62 @@ public:
return v_pos.template get<0>(vec_key.getKey());
}
+ /*! \brief Get the position of an element
+ *
+ * see the vector_dist iterator usage to get an element key
+ *
+ * \param vec_key element
+ *
+ * \return the position of the element in space
+ *
+ */
+ __device__ __host__ inline auto getPosOrig(int vec_key) -> decltype(v_pos.template get<0>(vec_key))
+ {
+ return v_pos.template get<0>(vec_key);
+ }
+
+ /*! \brief Get the position of an element
+ *
+ * see the vector_dist iterator usage to get an element key
+ *
+ * \param vec_key element
+ *
+ * \return the position of the element in space
+ *
+ */
+ __device__ __host__ inline auto getPosOrig(const vect_dist_key_dx & vec_key) -> decltype(v_pos.template get<0>(vec_key.getKey()))
+ {
+ return v_pos.template get<0>(vec_key.getKey());
+ }
+
+ /*! \brief Get the position of an element
+ *
+ * see the vector_dist iterator usage to get an element key
+ *
+ * \param vec_key element
+ *
+ * \return the position of the element in space
+ *
+ */
+ __device__ __host__ inline auto getPosOrig(int vec_key) const -> decltype(v_pos.template get<0>(vec_key))
+ {
+ return v_pos.template get<0>(vec_key);
+ }
+
+ /*! \brief Get the position of an element
+ *
+ * see the vector_dist iterator usage to get an element key
+ *
+ * \param vec_key element
+ *
+ * \return the position of the element in space
+ *
+ */
+ __device__ __host__ inline auto getPosOrig(const vect_dist_key_dx & vec_key) const -> decltype(v_pos.template get<0>(vec_key.getKey()))
+ {
+ return v_pos.template get<0>(vec_key.getKey());
+ }
+
/*! \brief Get the property of an element
*
* see the vector_dist iterator usage to get an element key
diff --git a/src/Vector/vector_dist_subset.hpp b/src/Vector/vector_dist_subset.hpp
index 0d25ea2c9..0e5a62d8b 100644
--- a/src/Vector/vector_dist_subset.hpp
+++ b/src/Vector/vector_dist_subset.hpp
@@ -269,6 +269,27 @@ public:
return vd.getPos(vect_dist_key_dx(pid.template get<0>(vec_key.getKey())));
}
+ /*! \brief Move the memory from the device to host memory
+ *
+ * \tparam property to move use POS_PROP for position property
+ *
+ */
+ template<unsigned int ... prp>
+ inline void hostToDeviceProp()
+ {
+ vd.template hostToDeviceProp<prp ...>();
+ }
+
+ /*! \brief Move the memory from the device to host memory
+ *
+ * \tparam property to move use POS_PROP for position property
+ *
+ */
+ inline void hostToDevicePos()
+ {
+ vd.template hostToDevicePos<0>();
+ }
+
/*! \brief Get the position of an element
*
* see the vector_dist iterator usage to get an element key
--
GitLab