Template/move CUDA texture cleanup code from NB CUDA module to cudautils.cu

[alexxy/gromacs.git] / src / gromacs / gpu_utils / cudautils.cuh

diff --git a/src/gromacs/gpu_utils/cudautils.cuh b/src/gromacs/gpu_utils/cudautils.cuh
index cfe5c2a381506ae3febaec07f84956bafcd1cd42..43b40a93fcd09af7c10bdcd3000be644fe48a93b 100644 (file)
--- a/src/gromacs/gpu_utils/cudautils.cuh
+++ b/src/gromacs/gpu_utils/cudautils.cuh
@@ -165,14 +165,8 @@ float cu_event_elapsed(cudaEvent_t /*start*/, cudaEvent_t /*end*/);
 /*! Waits for event end to complete and calculates the time between start and end. */
 int cu_wait_event_time(cudaEvent_t /*end*/, cudaEvent_t /*begin*/, float * /*time*/);
 
-/*! \brief Return whether texture objects are used on this device.
- *
- * \todo This should be static in cudautils.cu, as soon as texture destruction code is moved there as well
- *
- * \param[in]   pointer to the GPU device info structure to inspect for texture objects support
- * \return      true if texture objects are used on this device
- */
-bool use_texobj(const gmx_device_info_t *dev_info);
+// TODO: the 2 functions below are pretty much a constructor/destructor of a simple
+// GPU table object. We just need to add a templated __device__ table data fetching to complete it.
 
 /*! \brief Initialize parameter lookup table.
  *
@@ -195,4 +189,20 @@ void initParamLookupTable(T                        * &d_ptr,
                           int                        numElem,
                           const gmx_device_info_t   *devInfo);
 
+/*! \brief Destroy parameter lookup table.
+ *
+ * Unbinds texture reference/object, deallocates device memory.
+ *
+ * \tparam[in] T         Raw data type
+ * \param[in]  d_ptr     Device pointer to the memory to be deallocated
+ * \param[in]  texObj    Texture object to be deinitialized
+ * \param[in]  texRef    Texture reference to be deinitialized
+ * \param[in]  devInfo   Pointer to the info struct of the device in use
+ */
+template <typename T>
+void destroyParamLookupTable(T                         *d_ptr,
+                             cudaTextureObject_t        texObj,
+                             const struct texture<T, 1, cudaReadModeElementType> *texRef,
+                             const gmx_device_info_t   *devInfo);
+
 #endif