From abc80fcc2f65f33250cb9e63b5b1052838acc67c Mon Sep 17 00:00:00 2001
From: Saleel Kudchadker <Saleel.Kudchadker@amd.com>
Date: Thu, 29 Aug 2024 19:05:31 +0000
Subject: [PATCH] SWDEV-301667 - Improve kernel logging

Change-Id: I4b2b1950e3ab7124fd41af9a92a677c48d6da5eb
---
 rocclr/device/rocm/rocvirtual.cpp | 6 +++++-
 rocclr/platform/runtime.cpp       | 2 --
 2 files changed, 5 insertions(+), 3 deletions(-)

diff --git a/rocclr/device/rocm/rocvirtual.cpp b/rocclr/device/rocm/rocvirtual.cpp
index 62351cc1bb..7328642bb8 100644
--- a/rocclr/device/rocm/rocvirtual.cpp
+++ b/rocclr/device/rocm/rocvirtual.cpp
@@ -1439,6 +1439,8 @@ void* VirtualGPU::allocKernArg(size_t size, size_t alignment) {
     //! That means the app didn't call clFlush/clFinish for very long time.
     // Reset the signal for the barrier packet
     hsa_signal_silent_store_relaxed(kernarg_pool_signal_[active_chunk_], kInitSignalValueOne);
+    ClPrint(amd::LOG_INFO, amd::LOG_KERN, "Issue barrier to flush kernel arg chunk %d",
+            active_chunk_);
     // Dispatch a barrier packet into the queue
     dispatchBarrierPacket(kBarrierPacketHeader, true, kernarg_pool_signal_[active_chunk_]);
     // Get the next chunk
@@ -3265,7 +3267,9 @@ bool VirtualGPU::submitKernelInternal(const amd::NDRangeContainer& sizes,
                                                gpuKernel.KernargSegmentAlignment());
         currCmd_->SetKernelName(gpuKernel.name());
       } else {
-
+        ClPrint(amd::LOG_INFO, amd::LOG_KERN, "KernargSegmentByteSize = %lu "
+                "KernargSegmentAlignment = %lu", gpuKernel.KernargSegmentByteSize(),
+                gpuKernel.KernargSegmentAlignment());
         argBuffer = reinterpret_cast<address>(
             allocKernArg(gpuKernel.KernargSegmentByteSize(),
                          gpuKernel.KernargSegmentAlignment()));
diff --git a/rocclr/platform/runtime.cpp b/rocclr/platform/runtime.cpp
index 945f54527a..8efcdab216 100644
--- a/rocclr/platform/runtime.cpp
+++ b/rocclr/platform/runtime.cpp
@@ -82,7 +82,6 @@ bool Runtime::init() {
 
   initialized_ = true;
   pid_ = amd::Os::getProcessId();
-  ClTrace(LOG_DEBUG, LOG_INIT);
   return true;
 }
 
@@ -90,7 +89,6 @@ void Runtime::tearDown() {
   if (!initialized_) {
     return;
   }
-  ClTrace(LOG_DEBUG, LOG_INIT);
 
   Agent::tearDown();
   Device::tearDown();