From 99e6d330689a94d51709af7e777ac985bfbf362b Mon Sep 17 00:00:00 2001
From: Abhinav <absingh@mpi-cbg.de>
Date: Wed, 1 Mar 2023 02:23:05 +0100
Subject: [PATCH] Odeint GPU Support under construction.
---
src/Vector/cuda/map_vector_cuda_ker.cuh | 50 +++++++++++++++++++++++++
src/Vector/map_vector.hpp | 43 ++++++++++++++++-----
2 files changed, 83 insertions(+), 10 deletions(-)
diff --git a/src/Vector/cuda/map_vector_cuda_ker.cuh b/src/Vector/cuda/map_vector_cuda_ker.cuh
index ea202a7f..cbff5512 100644
--- a/src/Vector/cuda/map_vector_cuda_ker.cuh
+++ b/src/Vector/cuda/map_vector_cuda_ker.cuh
@@ -172,6 +172,11 @@ namespace openfpm
return v_size;
}
+ __host__ __device__ size_t size_local() const
+ {
+ return size();
+ }
+
/*! \brief return the maximum capacity of the vector before reallocation
*
* \return the capacity of the vector
@@ -204,6 +209,38 @@ namespace openfpm
return base.template get<p>(key);
}
+ /*! \brief Get an element of the vector
+ *
+ * Get an element of the vector
+ *
+ * \tparam p Property to get
+ * \param id Element to get
+ *
+ * \return the element value requested
+ *
+ */
+ template <unsigned int p>
+ __device__ __host__ inline auto getProp(unsigned int id) const -> decltype(base.template get<p>(grid_key_dx<1>(0)))
+ {
+ return this->get<p>(id);
+ }
+
+
+ /*! \brief Get an element of the vector
+ *
+ * Get an element of the vector
+ *
+ * \tparam p Property to get
+ * \param id Element to get
+ *
+ * \return the element value requested
+ *
+ */
+ template <unsigned int p, typename key_type>
+ __device__ __host__ inline auto getProp(key_type id) const -> decltype(base.template get<p>(grid_key_dx<1>(0)))
+ {
+ return this->get<p>(id.getKey());
+ }
/*! \brief Get an element of the vector
*
@@ -488,6 +525,14 @@ namespace openfpm
return base.getGPUIterator(start,stop,n_thr);
}
+ /*! \brief Get a domain iterator for the GPU
+ *
+ *
+ */
+ ite_gpu<1> getDomainIteratorGPU(size_t n_thr = default_kernel_wg_threads_) const
+ {
+ return getGPUIterator(n_thr);
+ }
/*! \brief Get an iterator for the GPU
*
@@ -596,6 +641,11 @@ namespace openfpm
return vref.size();
}
+ __host__ __device__ size_t size_local() const
+ {
+ return size();
+ }
+
__device__ __host__ unsigned int capacity() const
{
return vref.capacity;
diff --git a/src/Vector/map_vector.hpp b/src/Vector/map_vector.hpp
index 953fdeab..078ffd17 100644
--- a/src/Vector/map_vector.hpp
+++ b/src/Vector/map_vector.hpp
@@ -336,7 +336,7 @@ namespace openfpm
*
* \return the size
*
- */
+ */ //remove host device
size_t size_local() const
{
return v_size;
@@ -1359,7 +1359,21 @@ namespace openfpm
return base.get_o(key);
}
-
+ /*! \brief Get an element of the vector
+ *
+ * Get an element of the vector
+ *
+ * \tparam p Property to get
+ * \param id Element to get
+ *
+ * \return the element value requested
+ *
+ */ //remove device host
+ template <unsigned int p>
+ inline auto getProp(const unsigned int & id) -> decltype(base.template get<p>(grid_key_dx<1>(0)))
+ { //uncomment this
+ return this->template get<p>(id);
+ }
/*! \brief Get an element of the vector
*
* Get an element of the vector
@@ -1369,12 +1383,12 @@ namespace openfpm
*
* \return the element value requested
*
- */
-
- template <unsigned int p,typename KeyType>
- inline auto getProp(const KeyType & id) -> decltype(base.template get<p>(grid_key_dx<1>(0)))
+ *///remove host device
+ template <unsigned int p,typename KeyType>
+ inline auto getProp(const KeyType & id) -> decltype(base.template get<p>(grid_key_dx<1>(0)))
{
- return this->template get<p>(id.getKey());
+ //uncomment this
+ return this->template get<p>(id.getKey());
}
/*! \brief Get an element of the vector
@@ -1386,10 +1400,10 @@ namespace openfpm
*
* \return the element value requested
*
- */
+ */ //remove device host
template <unsigned int p, typename keyType>
- inline auto getProp(const keyType & id) const -> decltype(base.template get<p>(grid_key_dx<1>(0)))
- {
+ inline auto getProp(const keyType & id) const -> decltype(base.template get<p>(grid_key_dx<1>(0)))
+ { //uncomment this
return this->template get<p>(id.getKey());
}
@@ -1930,6 +1944,7 @@ namespace openfpm
return base.getGPUIterator(start,stop_,n_thr);
}
+
#endif
/*! \brief Get the vector elements iterator
@@ -1987,6 +2002,14 @@ namespace openfpm
return base.getGPUIterator(start,stop,n_thr);
}
+ /*! \brief Get a domain iterator for the GPU
+ *
+ *
+ */
+ ite_gpu<1> getDomainIteratorGPU(size_t n_thr = default_kernel_wg_threads_) const
+ {
+ return getGPUIterator(n_thr);
+ }
#endif
/*! \brief Return the size of the message needed to pack this object
--
GitLab