AI/rocm-systems
2
0
Garpu 0
Kode Masalah Tarik Permintaan Actions Packages Projects Rilis Wiki Kegiatan

SWDEV-485763 - Fix memory leaks in various unit tests

Telusuri Sumber 
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
This commit is contained in:
Marko Arandjelovic
2024-09-20 18:04:35 +02:00
orang tua 7bb49582db
melakukan 9cffda4ebb
33 mengubah file dengan 142 tambahan dan 38 penghapusan
Download Patch File Download Diff File Expand all files Collapse all files
catch/include/memcpy3d_tests_common.hh
+7 -1
Melihat File
@@ -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);
}
}
catch/include/resource_guards.hh
+7 -3
Melihat File
@@ -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<void>(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<void>(hipStreamDestroy(stream_));
}
}
@@ -361,7 +363,9 @@ class EventsGuard {
EventsGuard(EventsGuard&&) = delete;
~EventsGuard() {
for (auto& e : events_) static_cast<void>(hipEventDestroy(e));
for (auto& e : events_) {
static_cast<void>(hipEventDestroy(e));
}
}
hipEvent_t& operator[](int index) { return events_[index]; }
catch/unit/cooperativeGrps/hipCGMultiGridGroupType_old.cc
-1
Melihat File
@@ -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]));
}
}
catch/unit/graph/hipGraph.cc
+3
Melihat File
@@ -105,6 +105,9 @@ static void hipWithoutGraphs(float* inputVec_h, float* inputVec_d,
INFO("Time taken for hipWithoutGraphs : "
<< std::chrono::duration_cast<std::chrono::milliseconds>(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));
catch/unit/graph/hipGraphClone.cc
+1
Melihat File
@@ -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]) {
catch/unit/graph/hipGraphCloneComplx.cc
+6 -2
Melihat File
@@ -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));
}
catch/unit/graph/hipGraphCycle.cc
+5 -1
Melihat File
@@ -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));
}
/**
catch/unit/graph/hipGraphDebugDotPrint.cc
+1
Melihat File
@@ -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));
catch/unit/graph/hipGraphExecMemcpyNodeSetParamsFromSymbol_old.cc
+1
Melihat File
@@ -182,6 +182,7 @@ TEST_CASE("Unit_hipGraphExecMemcpyNodeSetParamsFromSymbol_Negative") {
}
HipTest::freeArrays<int>(A_d, B_d, nullptr,
A_h, B_h, nullptr, false);
HIP_CHECK(hipGraphExecDestroy(graphExec));
HIP_CHECK(hipGraphDestroy(graph));
}
catch/unit/graph/hipGraphExecMemcpyNodeSetParamsToSymbol.cc
+7 -7
Melihat File
@@ -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<int> 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<hipMemcpyKind>(-1)),
hipErrorInvalidValue);
}
HIP_CHECK(hipSetDevice(1));
LinearAllocGuard<int> 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<hipMemcpyKind>(-1)),
hipErrorInvalidValue);
}
HIP_CHECK(hipGraphExecDestroy(graph_exec));
catch/unit/graph/hipGraphExecMemsetNodeSetParams.cc
+6 -6
Melihat File
@@ -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<int> new_alloc(LinearAllocs::hipMalloc, 4 * sizeof(int));
params.dst = new_alloc.ptr();
HIP_CHECK_ERROR(hipGraphExecMemsetNodeSetParams(graph_exec, node, &params),
hipErrorInvalidValue);
}
HIP_CHECK(hipSetDevice(1));
LinearAllocGuard<int> new_alloc(LinearAllocs::hipMalloc, 4 * sizeof(int));
params.dst = new_alloc.ptr();
HIP_CHECK_ERROR(hipGraphExecMemsetNodeSetParams(graph_exec, node, &params),
hipErrorInvalidValue);
}
HIP_CHECK(hipGraphExecDestroy(graph_exec));
catch/unit/graph/hipGraphExecUpdate.cc
+2 -1
Melihat File
@@ -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);
catch/unit/graph/hipGraphPerf.cc
+8 -2
Melihat File
@@ -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<int>(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.
catch/unit/graph/hipGraphRemoveDependencies.cc
+3 -1
Melihat File
@@ -19,8 +19,9 @@ THE SOFTWARE.
#include <hip_test_common.hh>
#include <hip_test_checkers.hh>
#include <hip_test_kernels.hh>
#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));
}
catch/unit/graph/hipGraphRemoveDependencies_old.cc
+2 -1
Melihat File
@@ -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));
}
}
catch/unit/graph/hipSimpleGraphWithKernel.cc
+3
Melihat File
@@ -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));
}
/**
catch/unit/graph/hipStreamBeginCapture.cc
+1 -2
Melihat File
@@ -19,7 +19,7 @@ THE SOFTWARE.
#include <hip_test_common.hh>
#include <hip_test_kernels.hh>
#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));
catch/unit/graph/hipStreamBeginCaptureToGraph.cc
-1
Melihat File
@@ -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);
catch/unit/graph/hipStreamGetCaptureInfo_old.cc
+3 -1
Melihat File
@@ -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);
}
}
catch/unit/graph/hipStreamGetCaptureInfo_v2_old.cc
+2 -1
Melihat File
@@ -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));
}
}
catch/unit/graph/hipStreamIsCapturing_old.cc
+2 -1
Melihat File
@@ -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);
}
}
catch/unit/memory/hipMallocManaged.cc
+3
Melihat File
@@ -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);
}
catch/unit/memory/hipMallocMngdMultiThread.cc
+9
Melihat File
@@ -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));
}
catch/unit/memory/hipMemRangeGetAttributes_old.cc
+1
Melihat File
@@ -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];
catch/unit/memory/hipMemcpy2DAsync_old.cc
+1
Melihat File
@@ -418,6 +418,7 @@ TEST_CASE("Unit_hipMemcpy2DAsync_SizeCheck") {
// DeAllocating the memory
HIP_CHECK(hipFree(A_d));
HIP_CHECK(hipStreamDestroy(stream));
free(A_h);
}
catch/unit/memory/hipMemset.cc
+3
Melihat File
@@ -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));
}
catch/unit/module/hipExtModuleLaunchKernel.cc
+2
Melihat File
@@ -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));
}
}
catch/unit/multiThread/hipMultiThreadStreams2.cc
+7 -3
Melihat File
@@ -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]));
}
}
catch/unit/rtc/hipStreamCaptureRtc.cc
+1
Melihat File
@@ -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
catch/unit/stream/hipStreamCreateWithPriority.cc
+33 -2
Melihat File
@@ -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
catch/unit/stream/hipStreamGetDevice.cc
+1
Melihat File
@@ -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;
catch/unit/streamperthread/hipStreamPerThrdTsts.cc
+4
Melihat File
@@ -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;
catch/unit/streamperthread/hipStreamPerThread_Event.cc
+7 -1
Melihat File
@@ -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);
}
}
}
HIP_CHECK(hipHostFree(A_h));
HIP_CHECK(hipFree(A_d));
HIP_CHECK(hipEventDestroy(start));
HIP_CHECK(hipEventDestroy(end));
}