From 43e949746205410d5daade2b05db38a438d8ff61 Mon Sep 17 00:00:00 2001 From: jatang Date: Fri, 3 Feb 2023 17:42:01 -0500 Subject: [PATCH] SWDEV-380792 - Fix floating point exception when maxEngineClockFrequency_ is 0 Change-Id: Ic443ceae586c4c84995ed2abef9bd7f32f8b60f9 [ROCm/clr commit: b798c852720333e26d36d5dd3837953c8d165c61] --- projects/clr/rocclr/device/rocm/rocblit.cpp | 5 ++++- projects/clr/rocclr/device/rocm/rocdevice.cpp | 4 +++- 2 files changed, 7 insertions(+), 2 deletions(-) diff --git a/projects/clr/rocclr/device/rocm/rocblit.cpp b/projects/clr/rocclr/device/rocm/rocblit.cpp index a7a3efd26e..1bd6a0ff00 100644 --- a/projects/clr/rocclr/device/rocm/rocblit.cpp +++ b/projects/clr/rocclr/device/rocm/rocblit.cpp @@ -2691,7 +2691,10 @@ bool KernelBlitManager::runScheduler(uint64_t vqVM, amd::Memory* schedulerParam, sp->vqueue_header = vqVM; sp->parentAQL = sp->kernarg_address + sizeof(SchedulerParam); - sp->eng_clk = (1000 * 1024) / dev().info().maxEngineClockFrequency_; + + if (dev().info().maxEngineClockFrequency_ > 0) { + sp->eng_clk = (1000 * 1024) / dev().info().maxEngineClockFrequency_; + } // Use a device side global atomics to workaround the reliance of PCIe 3 atomics sp->write_index = hsa_queue_load_write_index_relaxed(schedulerQueue); diff --git a/projects/clr/rocclr/device/rocm/rocdevice.cpp b/projects/clr/rocclr/device/rocm/rocdevice.cpp index e63ab77eb0..e352577d02 100644 --- a/projects/clr/rocclr/device/rocm/rocdevice.cpp +++ b/projects/clr/rocclr/device/rocm/rocdevice.cpp @@ -1147,7 +1147,9 @@ bool Device::populateOCLDeviceConstants() { //TODO: add the assert statement for Raven if (!(isa().versionMajor() == 9 && isa().versionMinor() == 0 && isa().versionStepping() == 2)) { - assert(info_.maxEngineClockFrequency_ > 0); + if (info_.maxEngineClockFrequency_ <= 0) { + LogError("maxEngineClockFrequency_ is NOT positive!"); + } } if (HSA_STATUS_SUCCESS !=