From da5bff94647b7e2f2ddca85486aacc37798ef8f7 Mon Sep 17 00:00:00 2001 From: Satyanvesh Dittakavi Date: Thu, 11 Jul 2024 16:10:53 +0000 Subject: [PATCH] SWDEV-471935 - Destroy hsa queues with cumask set Fixes the memory leak with hipExtStreamCreateWithCUMask API. hsa queues with cumask set are not being reused and created everytime the API is called, But these queues were not being destroyed during hipStreamDestroy causing memory leak. Change-Id: Ibfbe019bbd73604e98eca80461efe53fa64bb701 [ROCm/clr commit: 191869b2521891df107e9e5cd4e2045c61f35abf] --- projects/clr/rocclr/device/rocm/rocdevice.cpp | 18 +++++++++++++++++- 1 file changed, 17 insertions(+), 1 deletion(-) diff --git a/projects/clr/rocclr/device/rocm/rocdevice.cpp b/projects/clr/rocclr/device/rocm/rocdevice.cpp index b9bd448b16..2ca56d354d 100644 --- a/projects/clr/rocclr/device/rocm/rocdevice.cpp +++ b/projects/clr/rocclr/device/rocm/rocdevice.cpp @@ -1,4 +1,4 @@ -/* Copyright (c) 2008 - 2022 Advanced Micro Devices, Inc. +/* Copyright (c) 2008 - 2024 Advanced Micro Devices, Inc. Permission is hereby granted, free of charge, to any person obtaining a copy of this software and associated documentation files (the "Software"), to deal @@ -3212,6 +3212,22 @@ void Device::releaseQueue(hsa_queue_t* queue, const std::vector& cuMas qInfo.refCount--; ClPrint(amd::LOG_INFO, amd::LOG_QUEUE, "releaseQueue refCount:%p (%d)", qIter->first->base_address, qIter->second.refCount); + // hsa queues with cumask set are not being reused. Hence, if the app uses multiple + // such queues it can cause memory leak and those must be destroyed here once the + // refcount reaches 0. + if ((!cuMask.empty()) && (qInfo.refCount == 0)) { + if (qInfo.hostcallBuffer_) { + ClPrint(amd::LOG_INFO, amd::LOG_QUEUE, + "Deleting hostcall buffer %p for hardware queue %p", + qInfo.hostcallBuffer_, qIter->first->base_address); + amd::disableHostcalls(qInfo.hostcallBuffer_); + context().svmFree(qInfo.hostcallBuffer_); + } + ClPrint(amd::LOG_INFO, amd::LOG_QUEUE, "Deleting hardware queue %p with refCount 0", + queue->base_address); + qIter = it.erase(qIter); + hsa_queue_destroy(queue); + } } } }