Deprecate HIP(hipStream_t, bool)

nliber · Sep 1, 2023 · 30a4da9 · 30a4da9
1 parent 7353bdf
commit 30a4da9
Show file tree

Hide file tree

Showing 4 changed files with 24 additions and 16 deletions.
diff --git a/core/src/HIP/Kokkos_HIP.cpp b/core/src/HIP/Kokkos_HIP.cpp
@@ -103,16 +103,20 @@ HIP::HIP()
       "HIP instance constructor");
 }
 
-HIP::HIP(hipStream_t const stream, bool manage_stream)
+HIP::HIP(hipStream_t const stream, Impl::ManageStream manage_stream)
     : m_space_instance(new Impl::HIPInternal, [](Impl::HIPInternal* ptr) {
         ptr->finalize();
         delete ptr;
       }) {
   Impl::HIPInternal::singleton().verify_is_initialized(
       "HIP instance constructor");
-  m_space_instance->initialize(stream, manage_stream);
+  m_space_instance->initialize(stream, static_cast<bool>(manage_stream));
 }
 
+KOKKOS_DEPRECATED HIP::HIP(hipStream_t const stream, bool manage_stream)
+    : HIP(stream,
+          manage_stream ? Impl::ManageStream::yes : Impl::ManageStream::no) {}
+
 void HIP::print_configuration(std::ostream& os, bool /*verbose*/) const {
   os << "Device Execution Space:\n";
   os << "  KOKKOS_ENABLE_HIP: yes\n";

diff --git a/core/src/HIP/Kokkos_HIP.hpp b/core/src/HIP/Kokkos_HIP.hpp
@@ -27,7 +27,8 @@
 namespace Kokkos {
 namespace Impl {
 class HIPInternal;
-}
+enum class ManageStream : bool { no, yes };
+}  // namespace Impl
 /// \class HIP
 /// \brief Kokkos device for multicore processors in the host memory space.
 class HIP {
@@ -47,7 +48,9 @@ class HIP {
   using scratch_memory_space = ScratchMemorySpace<HIP>;
 
   HIP();
-  HIP(hipStream_t stream, bool manage_stream = false);
+  HIP(hipStream_t stream,
+      Impl::ManageStream manage_stream = Impl::ManageStream::no);
+  KOKKOS_DEPRECATED HIP(hipStream_t stream, bool manage_stream);
 
   //@}
   //------------------------------------

diff --git a/core/src/HIP/Kokkos_HIP_Instance.cpp b/core/src/HIP/Kokkos_HIP_Instance.cpp
@@ -428,6 +428,16 @@ void hip_internal_error_throw(hipError_t e, const char *name, const char *file,
 
 //----------------------------------------------------------------------------
 
+void Kokkos::Impl::create_HIP_instances(std::vector<HIP> &instances) {
+  for (int s = 0; s < int(instances.size()); s++) {
+    hipStream_t stream;
+    KOKKOS_IMPL_HIP_SAFE_CALL(hipStreamCreate(&stream));
+    instances[s] = HIP(stream, ManageStream::yes);
+  }
+}
+
+//----------------------------------------------------------------------------
+
 namespace Kokkos {
 HIP::size_type HIP::detect_device_count() {
   int hipDevCount;

diff --git a/core/src/HIP/Kokkos_HIP_Instance.hpp b/core/src/HIP/Kokkos_HIP_Instance.hpp
@@ -150,6 +150,7 @@ class HIPInternal {
   void release_team_scratch_space(int scratch_pool_id);
 };
 
+void create_HIP_instances(std::vector<HIP> &instances);
 }  // namespace Impl
 
 namespace Experimental {
@@ -158,24 +159,14 @@ namespace Experimental {
 //   Customization point for backends
 //   Default behavior is to return the passed in instance
 
-namespace Impl {
-inline void create_HIP_instances(std::vector<HIP> &instances) {
-  for (int s = 0; s < int(instances.size()); s++) {
-    hipStream_t stream;
-    KOKKOS_IMPL_HIP_SAFE_CALL(hipStreamCreate(&stream));
-    instances[s] = HIP(stream, true);
-  }
-}
-}  // namespace Impl
-
 template <class... Args>
 std::vector<HIP> partition_space(const HIP &, Args...) {
   static_assert(
       (... && std::is_arithmetic_v<Args>),
       "Kokkos Error: partitioning arguments must be integers or floats");
 
   std::vector<HIP> instances(sizeof...(Args));
-  Impl::create_HIP_instances(instances);
+  Kokkos::Impl::create_HIP_instances(instances);
   return instances;
 }
 
@@ -188,7 +179,7 @@ std::vector<HIP> partition_space(const HIP &, std::vector<T> const &weights) {
   // We only care about the number of instances to create and ignore weights
   // otherwise.
   std::vector<HIP> instances(weights.size());
-  Impl::create_HIP_instances(instances);
+  Kokkos::Impl::create_HIP_instances(instances);
   return instances;
 }
 }  // namespace Experimental