From f2a7ea39278fd734ffc75d8704f1daf5ba29508b Mon Sep 17 00:00:00 2001 From: Saleel Kudchadker Date: Thu, 11 Mar 2021 12:52:05 -0800 Subject: [PATCH] SWDEV-271010 - Increase active wait time Increase wait time for active wait to 50us Change-Id: I8f269ab25ecc6775e655b9eb36fafc5f41a59c95 [ROCm/clr commit: d034c48405bcce8e7f6b85ca58836d94720648b3] --- projects/clr/rocclr/device/rocm/rocvirtual.cpp | 4 ++-- projects/clr/rocclr/device/rocm/rocvirtual.hpp | 1 + 2 files changed, 3 insertions(+), 2 deletions(-) diff --git a/projects/clr/rocclr/device/rocm/rocvirtual.cpp b/projects/clr/rocclr/device/rocm/rocvirtual.cpp index 2b58a4715c..06e3566917 100644 --- a/projects/clr/rocclr/device/rocm/rocvirtual.cpp +++ b/projects/clr/rocclr/device/rocm/rocvirtual.cpp @@ -388,8 +388,8 @@ hsa_signal_t* VirtualGPU::HwQueueTracker::WaitingSignal(HwQueueEngine engine) { // Early signal status check if (hsa_signal_load_relaxed(prof_signal->signal_) > 0) { const Settings& settings = gpu_.dev().settings(); - // Actively wait on CPU for 30 us to avoid extra overheads of signal tracking on GPU - if (!WaitForSignal(prof_signal->signal_)) { + // Actively wait on CPU for 50 us to avoid extra overheads of signal tracking on GPU + if (!WaitForSignal(prof_signal->signal_)) { if (settings.cpu_wait_for_signal_) { // Wait on CPU for completion if requested CpuWaitForSignal(prof_signal); diff --git a/projects/clr/rocclr/device/rocm/rocvirtual.hpp b/projects/clr/rocclr/device/rocm/rocvirtual.hpp index d77390092c..964bba5338 100644 --- a/projects/clr/rocclr/device/rocm/rocvirtual.hpp +++ b/projects/clr/rocclr/device/rocm/rocvirtual.hpp @@ -53,6 +53,7 @@ constexpr static hsa_signal_value_t kInitSignalValueOne = 1; // Timeouts for HSA signal wait constexpr static uint64_t kTimeout30us = 30000; +constexpr static uint64_t kTimeout50us = 50000; constexpr static uint64_t kUnlimitedWait = std::numeric_limits::max(); template