From d62f77307968e63a9c0b5c1688a2927a1410510c Mon Sep 17 00:00:00 2001 From: BrianHarrisonAMD <169072757+BrianHarrisonAMD@users.noreply.github.com> Date: Tue, 10 Sep 2024 10:50:49 -0600 Subject: [PATCH] Fix stream not being set when calling hipMemsetAsync (#3244) * Fix stream not being set when calling hipMemsetAsync * fix clang format issue * Fix missing handle for additional ZeroOutBuffer call in 6.2 --------- Co-authored-by: Jun Liu --- src/include/miopen/solver/implicitgemm_ck_util.hpp | 9 +++++---- 1 file changed, 5 insertions(+), 4 deletions(-) diff --git a/src/include/miopen/solver/implicitgemm_ck_util.hpp b/src/include/miopen/solver/implicitgemm_ck_util.hpp index 6967313720..5a0e9304b6 100644 --- a/src/include/miopen/solver/implicitgemm_ck_util.hpp +++ b/src/include/miopen/solver/implicitgemm_ck_util.hpp @@ -376,9 +376,10 @@ class TransposeInstance Run(handle, kernels, out_ptr, buf_handle.get()); } - void ZeroOutBuffer() + void ZeroOutBuffer(const Handle& handle) { - [[maybe_unused]] auto status = hipMemset(buf_handle.get(), 0, tensor_sz); + [[maybe_unused]] auto status = + hipMemsetAsync(buf_handle.get(), 0, tensor_sz, handle.GetStream()); assert(status == hipSuccess); } @@ -702,7 +703,7 @@ ConvSolution InitInvokerFactoryNCHW(const ExecutionContext& ctx, /// \todo: Will need SetTensor() to properly zero out non-packed tensors if(output_tr_inst.GetConvOperandTag() == internal::ConvOperandTag::Weights) { - output_tr_inst.ZeroOutBuffer(); + output_tr_inst.ZeroOutBuffer(handle); } std::array tr_ptrs = { @@ -845,7 +846,7 @@ ConvSolution InitInvokerFactoryWrwNCHW(const ExecutionContext& ctx, /// \todo: Will need SetTensor() to properly zero out non-packed tensors if(output_tr_inst.GetConvOperandTag() == internal::ConvOperandTag::Weights) { - output_tr_inst.ZeroOutBuffer(); + output_tr_inst.ZeroOutBuffer(handle); } std::array tr_ptrs = {