From e5c155d3f383793505ce28594ed2f49d45c303fa Mon Sep 17 00:00:00 2001 From: "systems-assistant[bot]" <221163467+systems-assistant[bot]@users.noreply.github.com> Date: Thu, 11 Sep 2025 01:16:21 +0200 Subject: [PATCH] SWDEV-547453 - Resolve memory leaks in module tests (#535) --- .../catch/unit/module/hipDrvLaunchKernelEx.cc | 5 +++ .../module/hipGetProcAddressModuleApis.cc | 34 +++++++++++-------- 2 files changed, 25 insertions(+), 14 deletions(-) diff --git a/projects/hip-tests/catch/unit/module/hipDrvLaunchKernelEx.cc b/projects/hip-tests/catch/unit/module/hipDrvLaunchKernelEx.cc index 935bd49337..45cd16461c 100644 --- a/projects/hip-tests/catch/unit/module/hipDrvLaunchKernelEx.cc +++ b/projects/hip-tests/catch/unit/module/hipDrvLaunchKernelEx.cc @@ -124,6 +124,9 @@ TEST_CASE("Unit_hipDrvLaunchKernelEx_NegTsts") { HIP_CHECK_ERROR(hipDrvLaunchKernelEx(&invalidConfig, function, kernelParams, NULL), hipErrorInvalidConfiguration); } + + HIP_CHECK(hipModuleUnload(module)); + HIP_CHECK(hipFree(d_output)); } bool runTestDrvLaunch(const char* testName, std::string kernelFunc, int totalThreads, int blockSize, @@ -281,6 +284,8 @@ TEST_CASE("Unit_hipDrvLaunchKernelEx_With_Different_Kernels") { int result = 0; HIP_CHECK(hipMemcpy(&result, devMem, sizeof(result), hipMemcpyDefault)); REQUIRE(result == 100); + + HIP_CHECK(hipFree(devMem)); } SECTION("Cooperative kernel with no arguments") { diff --git a/projects/hip-tests/catch/unit/module/hipGetProcAddressModuleApis.cc b/projects/hip-tests/catch/unit/module/hipGetProcAddressModuleApis.cc index eca139530f..7af86ef31d 100644 --- a/projects/hip-tests/catch/unit/module/hipGetProcAddressModuleApis.cc +++ b/projects/hip-tests/catch/unit/module/hipGetProcAddressModuleApis.cc @@ -334,6 +334,8 @@ TEST_CASE("Unit_hipGetProcAddress_ModuleApisCooperativeKernels") { return; } + HIP_CHECK(hipSetDevice(0)); + void* hipModuleLaunchCooperativeKernel_ptr = nullptr; void* hipModuleLaunchCooperativeKernelMultiDevice_ptr = nullptr; void* hipLaunchCooperativeKernel_ptr = nullptr; @@ -431,13 +433,13 @@ TEST_CASE("Unit_hipGetProcAddress_ModuleApisCooperativeKernels") { int deviceCount = 0; HIP_CHECK(hipGetDeviceCount(&deviceCount)); - hipModule_t* module = new hipModule_t[deviceCount]; - hipFunction_t* function = new hipFunction_t[deviceCount]; - hipStream_t* streamArr = new hipStream_t[deviceCount]; + auto module = std::make_unique(deviceCount); + auto function = std::make_unique(deviceCount); + auto stream_arr = std::make_unique(deviceCount); for (int i = 0; i < deviceCount; ++i) { HIP_CHECK(hipSetDevice(i)); - HIP_CHECK(hipStreamCreate(&streamArr[i])); + HIP_CHECK(hipStreamCreate(&stream_arr[i])); HIP_CHECK(hipModuleLoad(&module[i], "addKernel.code")); REQUIRE(module[i] != nullptr); @@ -460,7 +462,7 @@ TEST_CASE("Unit_hipGetProcAddress_ModuleApisCooperativeKernels") { params[i].blockDimZ = 1; params[i].kernelParams = nullptr; params[i].sharedMemBytes = 0; - params[i].hStream = streamArr[i]; + params[i].hStream = stream_arr[i]; } HIP_CHECK(dyn_hipModuleLaunchCooperativeKernelMultiDevice_ptr(params.data(), deviceCount, 0)); @@ -470,7 +472,7 @@ TEST_CASE("Unit_hipGetProcAddress_ModuleApisCooperativeKernels") { } for (int i = 0; i < deviceCount; ++i) { - HIP_CHECK(hipStreamDestroy(streamArr[i])); + HIP_CHECK(hipStreamDestroy(stream_arr[i])); HIP_CHECK(hipModuleUnload(module[i])); } } @@ -489,13 +491,15 @@ TEST_CASE("Unit_hipGetProcAddress_ModuleApisCooperativeKernels") { int deviceCount = 0; HIP_CHECK(hipGetDeviceCount(&deviceCount)); - hipStream_t* streamArr = new hipStream_t[deviceCount]; + auto stream_arr = std::make_unique(deviceCount); for (int i = 0; i < deviceCount; ++i) { HIP_CHECK(hipSetDevice(i)); - HIP_CHECK(hipStreamCreate(&streamArr[i])); + HIP_CHECK(hipStreamCreate(&stream_arr[i])); } + HIP_CHECK(hipSetDevice(0)); + std::vector params(deviceCount); for (int i = 0; i < deviceCount; ++i) { @@ -504,7 +508,7 @@ TEST_CASE("Unit_hipGetProcAddress_ModuleApisCooperativeKernels") { params[i].blockDim = {1, 1, 1}; params[i].args = nullptr; params[i].sharedMem = 0; - params[i].stream = streamArr[i]; + params[i].stream = stream_arr[i]; } HIP_CHECK(dyn_hipLaunchCooperativeKernelMultiDevice_ptr(params.data(), deviceCount, 0)); @@ -514,7 +518,7 @@ TEST_CASE("Unit_hipGetProcAddress_ModuleApisCooperativeKernels") { } for (int i = 0; i < deviceCount; ++i) { - HIP_CHECK(hipStreamDestroy(streamArr[i])); + HIP_CHECK(hipStreamDestroy(stream_arr[i])); } } @@ -523,13 +527,15 @@ TEST_CASE("Unit_hipGetProcAddress_ModuleApisCooperativeKernels") { int deviceCount = 0; HIP_CHECK(hipGetDeviceCount(&deviceCount)); - hipStream_t* streamArr = new hipStream_t[deviceCount]; + auto stream_arr = std::make_unique(deviceCount); for (int i = 0; i < deviceCount; ++i) { HIP_CHECK(hipSetDevice(i)); - HIP_CHECK(hipStreamCreate(&streamArr[i])); + HIP_CHECK(hipStreamCreate(&stream_arr[i])); } + HIP_CHECK(hipSetDevice(0)); + std::vector params(deviceCount); for (int i = 0; i < deviceCount; ++i) { @@ -538,7 +544,7 @@ TEST_CASE("Unit_hipGetProcAddress_ModuleApisCooperativeKernels") { params[i].blockDim = {1, 1, 1}; params[i].args = nullptr; params[i].sharedMem = 0; - params[i].stream = streamArr[i]; + params[i].stream = stream_arr[i]; } HIP_CHECK(dyn_hipExtLaunchMultiKernelMultiDevice_ptr(params.data(), deviceCount, 0)); @@ -548,7 +554,7 @@ TEST_CASE("Unit_hipGetProcAddress_ModuleApisCooperativeKernels") { } for (int i = 0; i < deviceCount; ++i) { - HIP_CHECK(hipStreamDestroy(streamArr[i])); + HIP_CHECK(hipStreamDestroy(stream_arr[i])); } }