From 9cffda4ebb5af5b8898e95ced339368643b749d0 Mon Sep 17 00:00:00 2001 From: Marko Arandjelovic Date: Fri, 20 Sep 2024 18:04:35 +0200 Subject: [PATCH] SWDEV-485763 - Fix memory leaks in various unit tests Fix memory leaks by adding missing destroy calls for events, streams, and graphs at the end of tests. Ensure that every test case executes destroy calls, regardless of whether it passes or fails. Change-Id: I814e35c528d90ed2abb34d77377f1a7fd3f1f11c --- catch/include/memcpy3d_tests_common.hh | 8 ++++- catch/include/resource_guards.hh | 10 ++++-- .../hipCGMultiGridGroupType_old.cc | 1 - catch/unit/graph/hipGraph.cc | 3 ++ catch/unit/graph/hipGraphClone.cc | 1 + catch/unit/graph/hipGraphCloneComplx.cc | 8 +++-- catch/unit/graph/hipGraphCycle.cc | 6 +++- catch/unit/graph/hipGraphDebugDotPrint.cc | 1 + ...phExecMemcpyNodeSetParamsFromSymbol_old.cc | 1 + ...hipGraphExecMemcpyNodeSetParamsToSymbol.cc | 14 ++++---- .../graph/hipGraphExecMemsetNodeSetParams.cc | 12 +++---- catch/unit/graph/hipGraphExecUpdate.cc | 3 +- catch/unit/graph/hipGraphPerf.cc | 10 ++++-- .../unit/graph/hipGraphRemoveDependencies.cc | 4 ++- .../graph/hipGraphRemoveDependencies_old.cc | 3 +- catch/unit/graph/hipSimpleGraphWithKernel.cc | 3 ++ catch/unit/graph/hipStreamBeginCapture.cc | 3 +- .../graph/hipStreamBeginCaptureToGraph.cc | 1 - .../unit/graph/hipStreamGetCaptureInfo_old.cc | 4 ++- .../graph/hipStreamGetCaptureInfo_v2_old.cc | 3 +- catch/unit/graph/hipStreamIsCapturing_old.cc | 3 +- catch/unit/memory/hipMallocManaged.cc | 3 ++ catch/unit/memory/hipMallocMngdMultiThread.cc | 9 +++++ .../memory/hipMemRangeGetAttributes_old.cc | 1 + catch/unit/memory/hipMemcpy2DAsync_old.cc | 1 + catch/unit/memory/hipMemset.cc | 3 ++ catch/unit/module/hipExtModuleLaunchKernel.cc | 2 ++ .../multiThread/hipMultiThreadStreams2.cc | 10 ++++-- catch/unit/rtc/hipStreamCaptureRtc.cc | 1 + .../stream/hipStreamCreateWithPriority.cc | 35 +++++++++++++++++-- catch/unit/stream/hipStreamGetDevice.cc | 1 + .../streamperthread/hipStreamPerThrdTsts.cc | 4 +++ .../hipStreamPerThread_Event.cc | 8 ++++- 33 files changed, 142 insertions(+), 38 deletions(-) diff --git a/catch/include/memcpy3d_tests_common.hh b/catch/include/memcpy3d_tests_common.hh index 58d5be6755..1b8c136e27 100644 --- a/catch/include/memcpy3d_tests_common.hh +++ b/catch/include/memcpy3d_tests_common.hh @@ -181,6 +181,9 @@ void Memcpy3DDeviceToDeviceShell(F memcpy_func, hipStream_t kernel_stream = null } if constexpr (enable_peer_access) { if (src_device == dst_device) { + if (device_count > 0 && kernel_stream != nullptr && kernel_stream != hipStreamPerThread) { + HIP_CHECK(hipStreamDestroy(kernel_stream)); + } return; } int can_access_peer = 0; @@ -189,6 +192,9 @@ void Memcpy3DDeviceToDeviceShell(F memcpy_func, hipStream_t kernel_stream = null std::string msg = "Skipped as peer access cannot be enabled between devices " + std::to_string(src_device) + " " + std::to_string(dst_device); HipTest::HIP_SKIP_TEST(msg.c_str()); + if (device_count > 0 && kernel_stream != nullptr && kernel_stream != hipStreamPerThread) { + HIP_CHECK(hipStreamDestroy(kernel_stream)); + } return; } HIP_CHECK(hipDeviceEnablePeerAccess(dst_device, 0)); @@ -888,4 +894,4 @@ void DrvMemcpy3DArrayDeviceShell(F memcpy_func, const hipStream_t kernel_stream }; PitchedMemoryVerify(host_alloc.ptr(), extent.width, extent.width / sizeof(int), extent.height, extent.depth, f); -} \ No newline at end of file +} diff --git a/catch/include/resource_guards.hh b/catch/include/resource_guards.hh index 91d8b6ffd5..c545fad074 100644 --- a/catch/include/resource_guards.hh +++ b/catch/include/resource_guards.hh @@ -307,8 +307,10 @@ class StreamGuard { break; case Streams::withFlags: HIP_CHECK(hipStreamCreateWithFlags(&stream_, flags_)); + break; case Streams::withPriority: HIP_CHECK(hipStreamCreateWithPriority(&stream_, flags_, priority_)); + break; } } @@ -318,7 +320,7 @@ class StreamGuard { StreamGuard& operator=(StreamGuard&& o) { if (this != &o) { - if (stream_type_ == Streams::created) { + if (stream_type_ >= Streams::created) { static_cast(hipStreamDestroy(stream_)); } @@ -337,7 +339,7 @@ class StreamGuard { } ~StreamGuard() { - if (stream_type_ == Streams::created && stream_ != nullptr) { + if (stream_type_ >= Streams::created && stream_ != nullptr) { static_cast(hipStreamDestroy(stream_)); } } @@ -361,7 +363,9 @@ class EventsGuard { EventsGuard(EventsGuard&&) = delete; ~EventsGuard() { - for (auto& e : events_) static_cast(hipEventDestroy(e)); + for (auto& e : events_) { + static_cast(hipEventDestroy(e)); + } } hipEvent_t& operator[](int index) { return events_[index]; } diff --git a/catch/unit/cooperativeGrps/hipCGMultiGridGroupType_old.cc b/catch/unit/cooperativeGrps/hipCGMultiGridGroupType_old.cc index a9a66dfe99..87b8f63017 100644 --- a/catch/unit/cooperativeGrps/hipCGMultiGridGroupType_old.cc +++ b/catch/unit/cooperativeGrps/hipCGMultiGridGroupType_old.cc @@ -374,7 +374,6 @@ static void test_cg_multi_grid_group_type(F kernel_func, int num_devices, int bl HIP_CHECK(hipHostFree(size_host[i])); HIP_CHECK(hipHostFree(thd_rank_host[i])); HIP_CHECK(hipHostFree(is_valid_host[i])); - HIP_CHECK(hipStreamDestroy(stream[i])); } } diff --git a/catch/unit/graph/hipGraph.cc b/catch/unit/graph/hipGraph.cc index 59c552c753..e1c1289584 100644 --- a/catch/unit/graph/hipGraph.cc +++ b/catch/unit/graph/hipGraph.cc @@ -105,6 +105,9 @@ static void hipWithoutGraphs(float* inputVec_h, float* inputVec_d, INFO("Time taken for hipWithoutGraphs : " << std::chrono::duration_cast(result).count() << " millisecs "); + HIP_CHECK(hipEventDestroy(forkStreamEvent)); + HIP_CHECK(hipEventDestroy(memsetEvent1)); + HIP_CHECK(hipEventDestroy(memsetEvent2)); HIP_CHECK(hipStreamDestroy(stream1)); HIP_CHECK(hipStreamDestroy(stream2)); HIP_CHECK(hipStreamDestroy(stream3)); diff --git a/catch/unit/graph/hipGraphClone.cc b/catch/unit/graph/hipGraphClone.cc index 75b83c82a0..14e418c193 100644 --- a/catch/unit/graph/hipGraphClone.cc +++ b/catch/unit/graph/hipGraphClone.cc @@ -208,6 +208,7 @@ void hipGraphClone_Func(bool ModifyOrigGraph = false) { HIP_CHECK(hipGraphInstantiate(&graphExec, graph, nullptr, nullptr, 0)); HIP_CHECK(hipGraphLaunch(graphExec, streamForGraph)); HIP_CHECK(hipStreamSynchronize(streamForGraph)); + HIP_CHECK(hipGraphExecDestroy(graphExec)); for (size_t i= 0; i < NElem; i++) { if (C_h[i] != B_h[i]) { diff --git a/catch/unit/graph/hipGraphCloneComplx.cc b/catch/unit/graph/hipGraphCloneComplx.cc index 6f0dd323ef..e8e40af506 100644 --- a/catch/unit/graph/hipGraphCloneComplx.cc +++ b/catch/unit/graph/hipGraphCloneComplx.cc @@ -1302,6 +1302,7 @@ static void hipGraphClone_Test_hipGraphEventRecordNodeSetEvent_and_Exec() { nullptr, nullptr, 0)); HIP_CHECK(hipGraphLaunch(clonedGraphExec, cg.stream)); HIP_CHECK(hipStreamSynchronize(cg.stream)); + HIP_CHECK(hipGraphExecDestroy(clonedGraphExec)); // Verify graph execution result HipTest::checkVectorADD(cg.B_h, cg.C_h, cg.A_h, N); @@ -1332,6 +1333,7 @@ static void hipGraphClone_Test_hipGraphEventRecordNodeSetEvent_and_Exec() { HipTest::checkVectorADD(cg.B_h, cg.C_h, cg.A_h, N); HIP_CHECK(hipEventDestroy(event_end2)); + HIP_CHECK(hipGraphExecDestroy(clonedGraphExec)); } SECTION("Verify hipGraphEventRecordNodeSetEvent & event_end->event_end3") { hipEvent_t event_end3; @@ -1414,9 +1416,9 @@ static void hipGraphClone_Test_hipGraphEventRecordNodeSetEvent_and_Exec() { HipTest::checkVectorADD(cg.B_h, cg.C_h, cg.A_h, N); HIP_CHECK(hipEventDestroy(event_end5)); + HIP_CHECK(hipGraphExecDestroy(clonedGraphExec)); } - HIP_CHECK(hipGraphExecDestroy(clonedGraphExec)); HIP_CHECK(hipGraphDestroy(clonedGraph)); HIP_CHECK(hipGraphDestroy(childgraph)); HIP_CHECK(hipEventDestroy(event_start)); @@ -1518,6 +1520,9 @@ static void hipGraphClone_Test_hipGraphEventWaitNodeSetEvent_and_Exec() { HIP_CHECK(hipGraphEventRecordNodeSetEvent(event_rec_node, event_2)); HIP_CHECK(hipGraphEventWaitNodeSetEvent(event_wait_node, event_2)); + // Destroy clonedGraphExec before instantating a new one + HIP_CHECK(hipGraphExecDestroy(clonedGraphExec)); + // Instantiate and launch the graph HIP_CHECK(hipGraphInstantiate(&clonedGraphExec, clonedGraph, nullptr, nullptr, 0)); @@ -2126,4 +2131,3 @@ TEST_CASE("Unit_hipGraphChild_hipUserObject_hipGraphUserObject") { HIP_CHECK(hipGraphDestroy(graph)); HIP_CHECK(hipStreamDestroy(stream)); } - diff --git a/catch/unit/graph/hipGraphCycle.cc b/catch/unit/graph/hipGraphCycle.cc index bb1a7e63ae..f11ea8c43b 100644 --- a/catch/unit/graph/hipGraphCycle.cc +++ b/catch/unit/graph/hipGraphCycle.cc @@ -56,6 +56,7 @@ TEST_CASE("Unit_hipGraph_BasicCyclic1") { HIP_CHECK(hipStreamCreate(&streamForGraph)); REQUIRE(hipErrorInvalidValue == hipGraphInstantiate(&graphExec, graph, nullptr, nullptr, 0)); HIP_CHECK(hipStreamDestroy(streamForGraph)); + HIP_CHECK(hipGraphDestroy(graph)); } /** @@ -83,13 +84,15 @@ TEST_CASE("Unit_hipGraph_BasicCyclic2") { HIP_CHECK(hipGraphRemoveDependencies(graph, &emptyNode3, &emptyNode1, 1)); HIP_CHECK(hipStreamCreate(&streamForGraph)); HIP_CHECK(hipGraphInstantiate(&graphExec, graph, nullptr, nullptr, 0)); + HIP_CHECK(hipGraphExecDestroy(graphExec)); + HIP_CHECK(hipGraphDestroy(graph)); HIP_CHECK(hipStreamDestroy(streamForGraph)); } /** * Tests basic functionality of cycle detection in hipGraph APIs by * Adding manual empty nodes - * Cyclic graph, cycle formation first, Remove edge causes disconnected graph which is still + * Cyclic graph, cycle formation first, Remove edge causes disconnected graph which is still * cyclic */ TEST_CASE("Unit_hipGraph_BasicCyclic3") { @@ -115,6 +118,7 @@ TEST_CASE("Unit_hipGraph_BasicCyclic3") { HIP_CHECK(hipStreamCreate(&streamForGraph)); REQUIRE(hipErrorInvalidValue == hipGraphInstantiate(&graphExec, graph, nullptr, nullptr, 0)); HIP_CHECK(hipStreamDestroy(streamForGraph)); + HIP_CHECK(hipGraphDestroy(graph)); } /** diff --git a/catch/unit/graph/hipGraphDebugDotPrint.cc b/catch/unit/graph/hipGraphDebugDotPrint.cc index 04e3c874e3..2ebf5ff741 100644 --- a/catch/unit/graph/hipGraphDebugDotPrint.cc +++ b/catch/unit/graph/hipGraphDebugDotPrint.cc @@ -284,6 +284,7 @@ static void hipGraphDebugDotPrint_Functional(const char* fName, free(mem_h); HIP_CHECK(hipFree(mem_d)); HipTest::freeArrays(A_d, B_d, C_d, A_h, B_h, C_h, false); + HIP_CHECK(hipEventDestroy(event)); HIP_CHECK(hipGraphExecDestroy(graphExec)); HIP_CHECK(hipGraphDestroy(graph)); HIP_CHECK(hipGraphDestroy(childGraph)); diff --git a/catch/unit/graph/hipGraphExecMemcpyNodeSetParamsFromSymbol_old.cc b/catch/unit/graph/hipGraphExecMemcpyNodeSetParamsFromSymbol_old.cc index b2a92f94db..a2a0fd1d28 100644 --- a/catch/unit/graph/hipGraphExecMemcpyNodeSetParamsFromSymbol_old.cc +++ b/catch/unit/graph/hipGraphExecMemcpyNodeSetParamsFromSymbol_old.cc @@ -182,6 +182,7 @@ TEST_CASE("Unit_hipGraphExecMemcpyNodeSetParamsFromSymbol_Negative") { } HipTest::freeArrays(A_d, B_d, nullptr, A_h, B_h, nullptr, false); + HIP_CHECK(hipGraphExecDestroy(graphExec)); HIP_CHECK(hipGraphDestroy(graph)); } diff --git a/catch/unit/graph/hipGraphExecMemcpyNodeSetParamsToSymbol.cc b/catch/unit/graph/hipGraphExecMemcpyNodeSetParamsToSymbol.cc index 4d3d0189d6..ee05f191b9 100644 --- a/catch/unit/graph/hipGraphExecMemcpyNodeSetParamsToSymbol.cc +++ b/catch/unit/graph/hipGraphExecMemcpyNodeSetParamsToSymbol.cc @@ -181,14 +181,14 @@ TEST_CASE("Unit_hipGraphExecMemcpyNodeSetParamsToSymbol_Negative_Parameters") { SECTION("Changing src allocation device") { if (HipTest::getDeviceCount() < 2) { HipTest::HIP_SKIP_TEST("Test requires two connected GPUs"); - return; + } else { + HIP_CHECK(hipSetDevice(1)); + LinearAllocGuard new_var(LinearAllocs::hipMalloc, sizeof(int)); + HIP_CHECK_ERROR(hipGraphExecMemcpyNodeSetParamsToSymbol( + graph_exec, node, SYMBOL(int_device_var), new_var.ptr(), + sizeof(*new_var.ptr()), 0, static_cast(-1)), + hipErrorInvalidValue); } - HIP_CHECK(hipSetDevice(1)); - LinearAllocGuard new_var(LinearAllocs::hipMalloc, sizeof(int)); - HIP_CHECK_ERROR(hipGraphExecMemcpyNodeSetParamsToSymbol( - graph_exec, node, SYMBOL(int_device_var), new_var.ptr(), - sizeof(*new_var.ptr()), 0, static_cast(-1)), - hipErrorInvalidValue); } HIP_CHECK(hipGraphExecDestroy(graph_exec)); diff --git a/catch/unit/graph/hipGraphExecMemsetNodeSetParams.cc b/catch/unit/graph/hipGraphExecMemsetNodeSetParams.cc index 8533ed4170..123838e80c 100644 --- a/catch/unit/graph/hipGraphExecMemsetNodeSetParams.cc +++ b/catch/unit/graph/hipGraphExecMemsetNodeSetParams.cc @@ -167,13 +167,13 @@ TEST_CASE("Unit_hipGraphExecMemsetNodeSetParams_Negative_Parameters") { SECTION("Changing dst allocation device") { if (HipTest::getDeviceCount() < 2) { HipTest::HIP_SKIP_TEST("Test requires two connected GPUs"); - return; + } else { + HIP_CHECK(hipSetDevice(1)); + LinearAllocGuard new_alloc(LinearAllocs::hipMalloc, 4 * sizeof(int)); + params.dst = new_alloc.ptr(); + HIP_CHECK_ERROR(hipGraphExecMemsetNodeSetParams(graph_exec, node, ¶ms), + hipErrorInvalidValue); } - HIP_CHECK(hipSetDevice(1)); - LinearAllocGuard new_alloc(LinearAllocs::hipMalloc, 4 * sizeof(int)); - params.dst = new_alloc.ptr(); - HIP_CHECK_ERROR(hipGraphExecMemsetNodeSetParams(graph_exec, node, ¶ms), - hipErrorInvalidValue); } HIP_CHECK(hipGraphExecDestroy(graph_exec)); diff --git a/catch/unit/graph/hipGraphExecUpdate.cc b/catch/unit/graph/hipGraphExecUpdate.cc index bbe00df59a..bbe91ae4b3 100644 --- a/catch/unit/graph/hipGraphExecUpdate.cc +++ b/catch/unit/graph/hipGraphExecUpdate.cc @@ -108,7 +108,6 @@ TEST_CASE("Unit_hipGraphExecUpdate_Negative_TypeChange") { dependencies.push_back(memsetNode); HIP_CHECK(hipGraphInstantiate(&graphExec, graph, nullptr, nullptr, 0)); HIP_CHECK(hipGraphCreate(&graph2, 0)); - HIP_CHECK(hipStreamCreate(&streamForGraph)); HIP_CHECK(hipGraphAddMemcpyNode1D(&memcpy_A, graph2, nullptr, 0, A_d, A_h, Nbytes, hipMemcpyHostToDevice)); // graphExec was created before memcpyTemp was added to graph. @@ -488,6 +487,7 @@ TEST_CASE("Unit_hipGraphExecUpdate_Negative_Functional_CountDiffer_2") { #else REQUIRE(hipSuccess == ret); #endif + HIP_CHECK(hipGraphExecDestroy(graphExec2)); HIP_CHECK(hipGraphDestroy(graph2)); } SECTION("When A node is deleted in GraphExec but not its pair from Graph") { @@ -506,6 +506,7 @@ TEST_CASE("Unit_hipGraphExecUpdate_Negative_Functional_CountDiffer_2") { #endif REQUIRE(NULL == hErrorNode_out); + HIP_CHECK(hipGraphExecDestroy(graphExec3)); HIP_CHECK(hipGraphDestroy(graph3)); } HipTest::freeArrays(A_d, B_d, C_d, A_h, B_h, C_h, false); diff --git a/catch/unit/graph/hipGraphPerf.cc b/catch/unit/graph/hipGraphPerf.cc index 3f7b2b1b5f..817030f1c0 100644 --- a/catch/unit/graph/hipGraphPerf.cc +++ b/catch/unit/graph/hipGraphPerf.cc @@ -386,6 +386,8 @@ static void checkGraphEventcontinuousKernelCall(const unsigned int kNumNode) { HipTest::checkVectorADD(A_h, B_h, C_h, N); HipTest::freeArrays(A_d, B_d, C_d, A_h, B_h, C_h, false); + HIP_CHECK(hipEventDestroy(eventstart)); + HIP_CHECK(hipEventDestroy(eventend)); HIP_CHECK(hipGraphExecDestroy(graphExec)); HIP_CHECK(hipGraphDestroy(graph)); HIP_CHECK(hipStreamDestroy(stream)); @@ -483,6 +485,9 @@ static void checkGraphEventcontinuousKernelCallIn2Blocks( HipTest::checkVectorSUB(A_h, B_h, C_h, N); HipTest::freeArrays(A_d, B_d, C_d, A_h, B_h, C_h, false); + HIP_CHECK(hipEventDestroy(eventstart)); + HIP_CHECK(hipEventDestroy(eventmid)); + HIP_CHECK(hipEventDestroy(eventend)); HIP_CHECK(hipGraphExecDestroy(graphExec)); HIP_CHECK(hipGraphDestroy(graph)); HIP_CHECK(hipStreamDestroy(stream)); @@ -582,6 +587,7 @@ static void hipGraph_PerfCheck_hipGraphExecKernelNodeSetParams( HIP_CHECK(hipGraphInstantiate(&graphExec, graph, NULL, NULL, 0)); HIP_CHECK(hipGraphLaunch(graphExec, stream)); HIP_CHECK(hipStreamSynchronize(stream)); + HIP_CHECK(hipGraphExecDestroy(graphExec)); // Verify graph execution result HipTest::checkVectorADD(A_h, B_h, C_h, N); @@ -2519,7 +2525,7 @@ static void hipGraph_PerfCheck_hipGraphExecUpdate(const hipStream_t& stream) { * ------------------------ * - Validate hipGraph performance with doorbell set. * - DEBUG_CLR_GRAPH_PACKET_CAPTURE - * 1) Added 2 nodes of MemCpy & a Kernel node and copy back result using memcpy + * 1) Added 2 nodes of MemCpy & a Kernel node and copy back result using memcpy and Instantiate graph & update new graph with similar node structure with api hipGraphExecUpdate and verify the result, the updated node should reflect. i) Check with Multi device case. @@ -2667,7 +2673,7 @@ static void hipGraph_PerfCheck_hipGraphExecUpdate_kernel_inLoop( * ------------------------ * - Validate hipGraph performance with doorbell set. * - DEBUG_CLR_GRAPH_PACKET_CAPTURE - * 1) Added 2 nodes of MemCpy & a Kernel node in sequence and copy back result using memcpy + * 1) Added 2 nodes of MemCpy & a Kernel node in sequence and copy back result using memcpy and Instantiate graph & update new graph with similar node structure with api hipGraphExecUpdate and verify the result, the updated node should reflect. i) Check with Multi device case. diff --git a/catch/unit/graph/hipGraphRemoveDependencies.cc b/catch/unit/graph/hipGraphRemoveDependencies.cc index 2cbc28b749..44762cc9cc 100644 --- a/catch/unit/graph/hipGraphRemoveDependencies.cc +++ b/catch/unit/graph/hipGraphRemoveDependencies.cc @@ -19,8 +19,9 @@ THE SOFTWARE. #include #include + #include - + #include "graph_dependency_common.hh" @@ -307,6 +308,7 @@ TEST_CASE("Unit_hipGraphRemoveDependencies_Positive_ChangeComputeFunc") { REQUIRE(false == bMismatch); // Destroy HipTest::freeArrays(A_d, B_d, C_d, A_h, B_h, C_h, false); + HIP_CHECK(hipGraphExecDestroy(graphExec)); HIP_CHECK(hipGraphDestroy(graph)); HIP_CHECK(hipStreamDestroy(streamForGraph)); } diff --git a/catch/unit/graph/hipGraphRemoveDependencies_old.cc b/catch/unit/graph/hipGraphRemoveDependencies_old.cc index 18cfd728cd..e41dc9e072 100644 --- a/catch/unit/graph/hipGraphRemoveDependencies_old.cc +++ b/catch/unit/graph/hipGraphRemoveDependencies_old.cc @@ -372,6 +372,7 @@ TEST_CASE("Unit_hipGraphRemoveDependencies_ChangeComputeFunc") { REQUIRE(false == bMismatch); // Destroy HipTest::freeArrays(A_d, B_d, C_d, A_h, B_h, C_h, false); + HIP_CHECK(hipGraphExecDestroy(graphExec)); HIP_CHECK(hipGraphDestroy(graph)); HIP_CHECK(hipStreamDestroy(streamForGraph)); } @@ -459,4 +460,4 @@ TEST_CASE("Unit_hipGraphRemoveDependencies_Negative") { HIP_CHECK(hipGraphDestroy(graph)); HIP_CHECK(hipEventDestroy(event_end)); HIP_CHECK(hipEventDestroy(event_start)); -} \ No newline at end of file +} diff --git a/catch/unit/graph/hipSimpleGraphWithKernel.cc b/catch/unit/graph/hipSimpleGraphWithKernel.cc index 08f20cbe45..537791d4a9 100644 --- a/catch/unit/graph/hipSimpleGraphWithKernel.cc +++ b/catch/unit/graph/hipSimpleGraphWithKernel.cc @@ -94,6 +94,8 @@ static void hipTestWithGraph() { delete[] out_h; HIP_CHECK(hipFree(in_d)); HIP_CHECK(hipFree(out_d)); + HIP_CHECK(hipGraphExecDestroy(instance)); + HIP_CHECK(hipStreamDestroy(stream)); } static void hipTestWithoutGraph() { @@ -143,6 +145,7 @@ static void hipTestWithoutGraph() { delete[] out_h; HIP_CHECK(hipFree(in_d)); HIP_CHECK(hipFree(out_d)); + HIP_CHECK(hipStreamDestroy(stream)); } /** diff --git a/catch/unit/graph/hipStreamBeginCapture.cc b/catch/unit/graph/hipStreamBeginCapture.cc index 78204efac9..ad15f2f451 100644 --- a/catch/unit/graph/hipStreamBeginCapture.cc +++ b/catch/unit/graph/hipStreamBeginCapture.cc @@ -19,7 +19,7 @@ THE SOFTWARE. #include #include - + #include "stream_capture_common.hh" // NOLINT #pragma clang diagnostic ignored "-Wunused-variable" @@ -218,7 +218,6 @@ static void interStrmEventSyncCapture(const hipStream_t& stream1, const hipStrea EventsGuard events_guard(1); hipEvent_t event = events_guard[0]; - HIP_CHECK(hipEventCreate(&event)); HIP_CHECK(hipStreamBeginCapture(stream1, hipStreamCaptureModeGlobal)); HIP_CHECK(hipEventRecord(event, stream1)); HIP_CHECK(hipStreamWaitEvent(stream2, event, 0)); diff --git a/catch/unit/graph/hipStreamBeginCaptureToGraph.cc b/catch/unit/graph/hipStreamBeginCaptureToGraph.cc index 20b1d5b13c..4b56eae8a8 100644 --- a/catch/unit/graph/hipStreamBeginCaptureToGraph.cc +++ b/catch/unit/graph/hipStreamBeginCaptureToGraph.cc @@ -969,7 +969,6 @@ TEST_CASE("Unit_hipStreamBeginCaptureToGraph_StateTesting") { HIP_CHECK(hipStreamCreate(&stream2)); HIP_CHECK(hipEventCreate(&e)); hipStreamCaptureStatus captureStatus = hipStreamCaptureStatusNone; - HIP_CHECK(hipStreamCreate(&stream1)); HIP_CHECK(hipGraphCreate(&graph, 0)); HIP_CHECK(hipStreamIsCapturing(stream1, &captureStatus)); REQUIRE(captureStatus == hipStreamCaptureStatusNone); diff --git a/catch/unit/graph/hipStreamGetCaptureInfo_old.cc b/catch/unit/graph/hipStreamGetCaptureInfo_old.cc index c4b0b4b5b1..e6833716e4 100644 --- a/catch/unit/graph/hipStreamGetCaptureInfo_old.cc +++ b/catch/unit/graph/hipStreamGetCaptureInfo_old.cc @@ -149,6 +149,7 @@ void validateStreamCaptureInfo(hipStream_t mstream) { HIP_CHECK(hipStreamSynchronize(streamForLaunch)); + HIP_CHECK(hipGraphExecDestroy(graphExec)); HIP_CHECK(hipGraphDestroy(graph)); HIP_CHECK(hipStreamDestroy(streamForLaunch)); HIP_CHECK(hipStreamDestroy(stream1)); @@ -610,6 +611,7 @@ TEST_CASE("Unit_hipStreamGetCaptureInfo_Nullstream_CaptureInfo") { REQUIRE(C_h[i] == D_h[i]); } + HIP_CHECK(hipGraphExecDestroy(graphExec)); HIP_CHECK(hipGraphDestroy(graph)); HIP_CHECK(hipStreamDestroy(stream)); HIP_CHECK(hipStreamDestroy(streamForGraph)); @@ -618,4 +620,4 @@ TEST_CASE("Unit_hipStreamGetCaptureInfo_Nullstream_CaptureInfo") { free(A_h); free(C_h); free(D_h); -} \ No newline at end of file +} diff --git a/catch/unit/graph/hipStreamGetCaptureInfo_v2_old.cc b/catch/unit/graph/hipStreamGetCaptureInfo_v2_old.cc index 47a168b02c..3d833d4250 100644 --- a/catch/unit/graph/hipStreamGetCaptureInfo_v2_old.cc +++ b/catch/unit/graph/hipStreamGetCaptureInfo_v2_old.cc @@ -158,6 +158,7 @@ void validateStreamCaptureInfoV2(hipStream_t mstream) { HIP_CHECK(hipStreamSynchronize(streamForLaunch)); + HIP_CHECK(hipGraphExecDestroy(graphExec)); HIP_CHECK(hipGraphDestroy(graph)); HIP_CHECK(hipStreamDestroy(streamForLaunch)); HIP_CHECK(hipStreamDestroy(stream1)); @@ -277,4 +278,4 @@ TEST_CASE("Unit_hipStreamGetCaptureInfo_v2_ParamValidation") { HIP_CHECK(hipGraphDestroy(graph)); HIP_CHECK(hipStreamDestroy(stream)); HIP_CHECK(hipFree(A_d)); -} \ No newline at end of file +} diff --git a/catch/unit/graph/hipStreamIsCapturing_old.cc b/catch/unit/graph/hipStreamIsCapturing_old.cc index ba4634a394..d963d43f49 100644 --- a/catch/unit/graph/hipStreamIsCapturing_old.cc +++ b/catch/unit/graph/hipStreamIsCapturing_old.cc @@ -425,6 +425,7 @@ TEST_CASE("Unit_hipStreamIsCapturing_ChkNullStrmStatus") { D_h[i] = A_h[i] * A_h[i]; REQUIRE(C_h[i] == D_h[i]); } + HIP_CHECK(hipGraphExecDestroy(graphExec)); HIP_CHECK(hipGraphDestroy(graph)); HIP_CHECK(hipStreamDestroy(stream)); HIP_CHECK(hipStreamDestroy(streamForGraph)); @@ -433,4 +434,4 @@ TEST_CASE("Unit_hipStreamIsCapturing_ChkNullStrmStatus") { free(A_h); free(C_h); free(D_h); -} \ No newline at end of file +} diff --git a/catch/unit/memory/hipMallocManaged.cc b/catch/unit/memory/hipMallocManaged.cc index 1b9766ffa9..04f579f0fa 100644 --- a/catch/unit/memory/hipMallocManaged.cc +++ b/catch/unit/memory/hipMallocManaged.cc @@ -138,6 +138,9 @@ TEST_CASE("Unit_hipMallocManaged_Advanced") { } HIP_CHECK(hipFree(A)); HIP_CHECK(hipFree(B)); + HIP_CHECK(hipFree(C)); + HIP_CHECK(hipEventDestroy(event0)); + HIP_CHECK(hipEventDestroy(event1)); REQUIRE(maxError != 0.0f); } diff --git a/catch/unit/memory/hipMallocMngdMultiThread.cc b/catch/unit/memory/hipMallocMngdMultiThread.cc index a163e9a0ab..4482c5e352 100644 --- a/catch/unit/memory/hipMallocMngdMultiThread.cc +++ b/catch/unit/memory/hipMallocMngdMultiThread.cc @@ -119,6 +119,9 @@ static void LaunchKrnl2(int* Hmm, size_t NumElms, int InitVal, int HmmMem) { INFO("Data Mismatch observed at line: " << __LINE__); REQUIRE(false); } + + HIP_CHECK(hipFree(ptr)); + HIP_CHECK(hipStreamDestroy(strm)); } static void LaunchKrnl3(int* Dptr, size_t NumElms, int InitVal) { @@ -143,6 +146,9 @@ static void LaunchKrnl3(int* Dptr, size_t NumElms, int InitVal) { INFO("Data Mismatch observed at line: " << __LINE__); REQUIRE(false); } + + HIP_CHECK(hipFree(Hmm)); + HIP_CHECK(hipStreamDestroy(strm)); } @@ -180,6 +186,9 @@ static void LaunchKrnl5(int* Hmm1, size_t NumElms, int InitVal, int KerneltoLaun INFO("Data Mismatch observed at line: " << __LINE__); REQUIRE(false); } + + HIP_CHECK(hipFree(Hmm2)); + HIP_CHECK(hipStreamDestroy(strm)); } diff --git a/catch/unit/memory/hipMemRangeGetAttributes_old.cc b/catch/unit/memory/hipMemRangeGetAttributes_old.cc index 63952f1bfd..7ca0f87c17 100644 --- a/catch/unit/memory/hipMemRangeGetAttributes_old.cc +++ b/catch/unit/memory/hipMemRangeGetAttributes_old.cc @@ -156,6 +156,7 @@ TEST_CASE("Unit_hipMemRangeGetAttributes_TstFlgs") { } HIP_CHECK(hipFree(Hmm)); + HIP_CHECK(hipStreamDestroy(strm)); delete[] AcsdBy; for (int i = 0; i < 4; ++i) { delete Outpt[i]; diff --git a/catch/unit/memory/hipMemcpy2DAsync_old.cc b/catch/unit/memory/hipMemcpy2DAsync_old.cc index 233d6077dd..69f8c73b6a 100644 --- a/catch/unit/memory/hipMemcpy2DAsync_old.cc +++ b/catch/unit/memory/hipMemcpy2DAsync_old.cc @@ -418,6 +418,7 @@ TEST_CASE("Unit_hipMemcpy2DAsync_SizeCheck") { // DeAllocating the memory HIP_CHECK(hipFree(A_d)); + HIP_CHECK(hipStreamDestroy(stream)); free(A_h); } diff --git a/catch/unit/memory/hipMemset.cc b/catch/unit/memory/hipMemset.cc index c4adb15e53..20bde29041 100644 --- a/catch/unit/memory/hipMemset.cc +++ b/catch/unit/memory/hipMemset.cc @@ -278,4 +278,7 @@ TEST_CASE("Unit_hipMemset_2AsyncOperations") { REQUIRE(v[0] == 0); REQUIRE(v[1024] == 1.75f); + + HIP_CHECK(hipFree(p2)); + HIP_CHECK(hipStreamDestroy(s)); } diff --git a/catch/unit/module/hipExtModuleLaunchKernel.cc b/catch/unit/module/hipExtModuleLaunchKernel.cc index c1d64b7388..2b3e9bc447 100644 --- a/catch/unit/module/hipExtModuleLaunchKernel.cc +++ b/catch/unit/module/hipExtModuleLaunchKernel.cc @@ -254,6 +254,7 @@ TEST_CASE("Unit_hipExtModuleLaunchKernel_Positive_Parameters") { start_event, nullptr)); HIP_CHECK(hipDeviceSynchronize()); HIP_CHECK(hipEventQuery(start_event)); + HIP_CHECK(hipEventDestroy(start_event)); } SECTION("Pass only stop event") { @@ -265,6 +266,7 @@ TEST_CASE("Unit_hipExtModuleLaunchKernel_Positive_Parameters") { nullptr, stop_event)); HIP_CHECK(hipDeviceSynchronize()); HIP_CHECK(hipEventQuery(stop_event)); + HIP_CHECK(hipEventDestroy(stop_event)); } } diff --git a/catch/unit/multiThread/hipMultiThreadStreams2.cc b/catch/unit/multiThread/hipMultiThreadStreams2.cc index d38754724d..c0968d0f37 100644 --- a/catch/unit/multiThread/hipMultiThreadStreams2.cc +++ b/catch/unit/multiThread/hipMultiThreadStreams2.cc @@ -51,7 +51,7 @@ void run1(size_t size, hipStream_t stream) { HIPCHECK(hipMemcpyAsync(Bh, Ah, size, hipMemcpyHostToHost, stream)); HIPCHECK(hipMemcpyAsync(Cd, Bh, size, hipMemcpyHostToDevice, stream)); hipLaunchKernelGGL(HIP_KERNEL_NAME(Inc), dim3(N / 500), dim3(500), 0, stream, Cd); - HIP_CHECK(hipGetLastError()); + HIP_CHECK(hipGetLastError()); HIPCHECK(hipMemcpyAsync(Dd, Cd, size, hipMemcpyDeviceToDevice, stream)); HIPCHECK(hipMemcpyAsync(Eh, Dd, size, hipMemcpyDeviceToHost, stream)); HIPCHECK(hipDeviceSynchronize()); @@ -93,9 +93,9 @@ void run(size_t size, hipStream_t stream1, hipStream_t stream2) { HIPCHECK(hipMemcpyAsync(Cd, Bh, size, hipMemcpyHostToDevice, stream1)); HIPCHECK(hipMemcpyAsync(Cdd, Bhh, size, hipMemcpyHostToDevice, stream2)); hipLaunchKernelGGL(HIP_KERNEL_NAME(Inc), dim3(N / 500), dim3(500), 0, stream1, Cd); - HIP_CHECK(hipGetLastError()); + HIP_CHECK(hipGetLastError()); hipLaunchKernelGGL(HIP_KERNEL_NAME(Inc), dim3(N / 500), dim3(500), 0, stream2, Cdd); - HIP_CHECK(hipGetLastError()); + HIP_CHECK(hipGetLastError()); HIPCHECK(hipMemcpyAsync(Dd, Cd, size, hipMemcpyDeviceToDevice, stream1)); HIPCHECK(hipMemcpyAsync(Ddd, Cdd, size, hipMemcpyDeviceToDevice, stream2)); HIPCHECK(hipMemcpyAsync(Eh, Dd, size, hipMemcpyDeviceToHost, stream1)); @@ -140,4 +140,8 @@ TEST_CASE("Unit_hipMultiThreadStreams2") { t2.join(); t3.join(); } + + for (int i = 0; i < 3; i++) { + HIPCHECK(hipStreamDestroy(stream[i])); + } } diff --git a/catch/unit/rtc/hipStreamCaptureRtc.cc b/catch/unit/rtc/hipStreamCaptureRtc.cc index ab5296c4d5..2264b4fa57 100644 --- a/catch/unit/rtc/hipStreamCaptureRtc.cc +++ b/catch/unit/rtc/hipStreamCaptureRtc.cc @@ -123,6 +123,7 @@ TEST_CASE("Unit_hipStreamCaptureRtc") { HIPCHECK(hipGraphLaunch(graph_exec, stream)); HIPCHECK(hipStreamSynchronize(stream)); + HIPCHECK(hipGraphExecDestroy(graph_exec)); HIPCHECK(hipStreamDestroy(stream)); // Check that the work was done diff --git a/catch/unit/stream/hipStreamCreateWithPriority.cc b/catch/unit/stream/hipStreamCreateWithPriority.cc index 7a246dd84e..37370bf500 100644 --- a/catch/unit/stream/hipStreamCreateWithPriority.cc +++ b/catch/unit/stream/hipStreamCreateWithPriority.cc @@ -488,12 +488,16 @@ bool validateStreamPrioritiesWithEvents() { OP(low, high) #undef OP - // free host & device memory + // free host & device memory & events #define OP(x) \ free(src_h_##x); \ free(dst_h_##x); \ HIP_CHECK(hipFree(src_d_##x)); \ - HIP_CHECK(hipFree(dst_d_##x)); + HIP_CHECK(hipFree(dst_d_##x)); \ + if (enable_priority_##x) { \ + HIP_CHECK(hipEventDestroy(event_start_##x)); \ + HIP_CHECK(hipEventDestroy(event_end_##x)); \ + } OP(low) OP(normal) OP(high) @@ -801,6 +805,33 @@ void TestForMultipleStreamWithPriority(void) { REQUIRE(memcmp(dst_h_high[i], src_h_high[i], size) == 0); } } + + for (int i = 0; i < LOW_PRIORITY_STREAMCOUNT; i++) { + if (enable_priority_low) { + HIP_CHECK(hipEventDestroy(event_start_low[i])); + HIP_CHECK(hipEventDestroy(event_end_low[i])); + } + + HIP_CHECK(hipStreamDestroy(stream_low[i])); + } + + for (int i = 0; i < NORMAL_PRIORITY_STREAMCOUNT; i++) { + if (enable_priority_normal) { + HIP_CHECK(hipEventDestroy(event_start_normal[i])); + HIP_CHECK(hipEventDestroy(event_end_normal[i])); + } + + HIP_CHECK(hipStreamDestroy(stream_normal[i])); + } + + for (int i = 0; i < HIGH_PRIORITY_STREAMCOUNT; i++) { + if (enable_priority_high) { + HIP_CHECK(hipEventDestroy(event_start_high[i])); + HIP_CHECK(hipEventDestroy(event_end_high[i])); + } + + HIP_CHECK(hipStreamDestroy(stream_high[i])); + } } } // namespace hipStreamCreateWithPriorityTest diff --git a/catch/unit/stream/hipStreamGetDevice.cc b/catch/unit/stream/hipStreamGetDevice.cc index fbe8324270..930fc1a0ea 100644 --- a/catch/unit/stream/hipStreamGetDevice.cc +++ b/catch/unit/stream/hipStreamGetDevice.cc @@ -140,6 +140,7 @@ static bool validateStreamGetDevice() { hipStream_t stream; HIP_CHECK(hipStreamCreate(&stream)); HIP_CHECK(hipStreamGetDevice(stream, &device_from_stream)); + HIP_CHECK(hipStreamDestroy(stream)); REQUIRE(device_from_stream == gpu); return true; diff --git a/catch/unit/streamperthread/hipStreamPerThrdTsts.cc b/catch/unit/streamperthread/hipStreamPerThrdTsts.cc index f940bf1766..d03a7dbae1 100644 --- a/catch/unit/streamperthread/hipStreamPerThrdTsts.cc +++ b/catch/unit/streamperthread/hipStreamPerThrdTsts.cc @@ -251,6 +251,9 @@ static void EventSync() { } else { IfTestPassed = true; } + + HIP_CHECK(hipEventDestroy(start)); + HIP_CHECK(hipEventDestroy(end)); } /* Launch a kernel in hipStreamPerThread, while it is in flight check for @@ -475,6 +478,7 @@ TEST_CASE("Unit_hipStreamPerThread_StrmWaitEvt") { } HIP_CHECK(hipFree(Ad)); HIP_CHECK(hipFree(Ad1)); + HIP_CHECK(hipEventDestroy(e1)); HIP_CHECK(hipStreamDestroy(Strm)); delete[] Ah; delete Ah1; diff --git a/catch/unit/streamperthread/hipStreamPerThread_Event.cc b/catch/unit/streamperthread/hipStreamPerThread_Event.cc index 6fef9914eb..1e721f2757 100644 --- a/catch/unit/streamperthread/hipStreamPerThread_Event.cc +++ b/catch/unit/streamperthread/hipStreamPerThread_Event.cc @@ -23,6 +23,7 @@ TEST_CASE("Unit_hipStreamPerThread_EventRecord") { hipEvent_t event; HIP_CHECK(hipEventCreate(&event)); HIP_CHECK(hipEventRecord(event, hipStreamPerThread)); + HIP_CHECK(hipEventDestroy(event)); } __global__ void update_even_odd(unsigned int N, int* out) { @@ -61,4 +62,9 @@ TEST_CASE("Unit_hipStreamPerThread_EventSynchronize") { REQUIRE(false); } } -} \ No newline at end of file + + HIP_CHECK(hipHostFree(A_h)); + HIP_CHECK(hipFree(A_d)); + HIP_CHECK(hipEventDestroy(start)); + HIP_CHECK(hipEventDestroy(end)); +}