Drop 2-arguments ZeroMemset constructor overloads (kokkos#6764)

* contiguous_fill_or_memset should pass a default-constructed exec space to ZeroMemset

* Rely on CTAD with ZeroMemset

No one wants to see these template parameters

* Drop 2-arguments ZeroMemset constructor overloads

* Make sure to include what we use (<cstring> for std::memset)

* Fix unused variable warning

core/src/Cuda/Kokkos_Cuda_ZeroMemset.hpp

@@ -33,15 +33,6 @@ struct ZeroMemset<Kokkos::Cuda, View<T, P...>> {
                  dst.data(), 0,
                  dst.size() * sizeof(typename View<T, P...>::value_type))));
   }
-
-  ZeroMemset(const View<T, P...>& dst,
-             typename View<T, P...>::const_value_type&) {
-    // FIXME_CUDA_MULTIPLE_DEVICES
-    KOKKOS_IMPL_CUDA_SAFE_CALL(
-        (Kokkos::Impl::CudaInternal::singleton().cuda_memset_wrapper(
-            dst.data(), 0,
-            dst.size() * sizeof(typename View<T, P...>::value_type))));
-  }
 };
 
 }  // namespace Impl

core/src/HIP/Kokkos_HIP_ZeroMemset.hpp

@@ -31,13 +31,6 @@ struct ZeroMemset<HIP, View<T, P...>> {
         dst.data(), 0, dst.size() * sizeof(typename View<T, P...>::value_type),
         exec_space.hip_stream()));
   }
-
-  ZeroMemset(const View<T, P...>& dst,
-             typename View<T, P...>::const_value_type&) {
-    KOKKOS_IMPL_HIP_SAFE_CALL(
-        hipMemset(dst.data(), 0,
-                  dst.size() * sizeof(typename View<T, P...>::value_type)));
-  }
 };
 
 }  // namespace Impl

core/src/Kokkos_CopyViews.hpp

@@ -1360,7 +1360,7 @@ contiguous_fill_or_memset(
       && !std::is_same_v<ExecutionSpace, Kokkos::OpenMP>
 #endif
   )
-    ZeroMemset<ExecutionSpace, View<DT, DP...>>(exec_space, dst, value);
+    ZeroMemset(exec_space, dst, value);
   else
     contiguous_fill(exec_space, dst, value);
 }
@@ -1386,15 +1386,16 @@ contiguous_fill_or_memset(
     typename ViewTraits<DT, DP...>::const_value_type& value) {
   using ViewType        = View<DT, DP...>;
   using exec_space_type = typename ViewType::execution_space;
+  exec_space_type exec;
 
 // On A64FX memset seems to do the wrong thing with regards to first touch
 // leading to the significant performance issues
 #ifndef KOKKOS_ARCH_A64FX
   if (Impl::is_zero_byte(value))
-    ZeroMemset<exec_space_type, View<DT, DP...>>(dst, value);
+    ZeroMemset(exec, dst, value);
   else
 #endif
-    contiguous_fill(exec_space_type(), dst, value);
+    contiguous_fill(exec, dst, value);
 }
 
 template <class DT, class... DP>

core/src/SYCL/Kokkos_SYCL_ZeroMemset.hpp

@@ -35,12 +35,6 @@ struct ZeroMemset<Kokkos::Experimental::SYCL, View<T, P...>> {
         ->m_queue->ext_oneapi_submit_barrier(std::vector<sycl::event>{event});
 #endif
   }
-
-  ZeroMemset(const View<T, P...>& dst,
-             typename View<T, P...>::const_value_type&) {
-    Experimental::Impl::SYCLInternal::singleton().m_queue->memset(
-        dst.data(), 0, dst.size() * sizeof(typename View<T, P...>::value_type));
-  }
 };
 
 }  // namespace Impl

core/src/Serial/Kokkos_Serial_ZeroMemset.hpp

@@ -22,6 +22,7 @@
 #include <Serial/Kokkos_Serial.hpp>
 
 #include <type_traits>
+#include <cstring>
 
 namespace Kokkos {
 namespace Impl {
@@ -34,14 +35,12 @@ template <class T, class... P>
 struct ZeroMemset<
     std::conditional_t<!std::is_same<Serial, DefaultHostExecutionSpace>::value,
                        Serial, DummyExecutionSpace>,
-    View<T, P...>>
-    : public ZeroMemset<DefaultHostExecutionSpace, View<T, P...>> {
-  using Base = ZeroMemset<DefaultHostExecutionSpace, View<T, P...>>;
-  using Base::Base;
-
+    View<T, P...>> {
   ZeroMemset(const Serial&, const View<T, P...>& dst,
-             typename View<T, P...>::const_value_type& value)
-      : Base(dst, value) {}
+             typename View<T, P...>::const_value_type&) {
+    using ValueType = typename View<T, P...>::value_type;
+    std::memset(dst.data(), 0, sizeof(ValueType) * dst.size());
+  }
 };
 
 }  // namespace Impl

core/src/impl/Kokkos_HostSpace_ZeroMemset.hpp

@@ -36,12 +36,6 @@ struct ZeroMemset<HostSpace::execution_space, View<T, P...>> {
     using ValueType = typename View<T, P...>::value_type;
     std::memset(dst.data(), 0, sizeof(ValueType) * dst.size());
   }
-
-  ZeroMemset(const View<T, P...>& dst,
-             typename View<T, P...>::const_value_type&) {
-    using ValueType = typename View<T, P...>::value_type;
-    std::memset(dst.data(), 0, sizeof(ValueType) * dst.size());
-  }
 };
 
 }  // end namespace Impl

core/src/impl/Kokkos_ViewMapping.hpp

@@ -2917,9 +2917,7 @@ struct ViewValueFunctor<DeviceType, ValueType, false /* is_scalar */> {
             "Kokkos::View::initialization [" + name + "] via memset",
             Kokkos::Profiling::Experimental::device_id(space), &kpID);
       }
-      (void)ZeroMemset<
-          ExecSpace, Kokkos::View<ValueType*, typename DeviceType::memory_space,
-                                  Kokkos::MemoryTraits<Kokkos::Unmanaged>>>(
+      (void)ZeroMemset(
           space,
           Kokkos::View<ValueType*, typename DeviceType::memory_space,
                        Kokkos::MemoryTraits<Kokkos::Unmanaged>>(ptr, n),
@@ -3051,9 +3049,7 @@ struct ViewValueFunctor<DeviceType, ValueType, true /* is_scalar */> {
             Kokkos::Profiling::Experimental::device_id(space), &kpID);
       }
 
-      (void)ZeroMemset<
-          ExecSpace, Kokkos::View<ValueType*, typename DeviceType::memory_space,
-                                  Kokkos::MemoryTraits<Kokkos::Unmanaged>>>(
+      (void)ZeroMemset(
           space,
           Kokkos::View<ValueType*, typename DeviceType::memory_space,
                        Kokkos::MemoryTraits<Kokkos::Unmanaged>>(ptr, n),