diff --git a/projects/hip-tests/catch/unit/warp/hipMatchSyncAllTests.cc b/projects/hip-tests/catch/unit/warp/hipMatchSyncAllTests.cc index dca3161a96..23026c301f 100644 --- a/projects/hip-tests/catch/unit/warp/hipMatchSyncAllTests.cc +++ b/projects/hip-tests/catch/unit/warp/hipMatchSyncAllTests.cc @@ -67,6 +67,10 @@ template static void runTestMatchAll_1() { for (int i = 0; i != warpSize; ++i) { REQUIRE(Predicate[i] == ExpPredicate[i]); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); + HIP_CHECK(hipFree(d_Predicate)); } template @@ -123,6 +127,10 @@ template static void runTestMatchAll_2() { for (int i = 0; i != warpSize; ++i) { REQUIRE(Predicate[i] == ExpPredicate[i]); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); + HIP_CHECK(hipFree(d_Predicate)); } template @@ -242,6 +250,10 @@ template static void runTestMatchAll_3() { for (int i = 0; i != warpSize; ++i) { REQUIRE(Predicate[i] == ExpPredicate[i]); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); + HIP_CHECK(hipFree(d_Predicate)); } __global__ void matchAll_4(int* Input, int* Output) { @@ -291,6 +303,9 @@ static void runTestMatchAll_4() { for (size_t i = 0; i < Output.size(); i++) { REQUIRE(Output[i] == Expected[i]); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } /** diff --git a/projects/hip-tests/catch/unit/warp/hipMatchSyncAnyTests.cc b/projects/hip-tests/catch/unit/warp/hipMatchSyncAnyTests.cc index 4973300bd2..5896e7b95c 100644 --- a/projects/hip-tests/catch/unit/warp/hipMatchSyncAnyTests.cc +++ b/projects/hip-tests/catch/unit/warp/hipMatchSyncAnyTests.cc @@ -67,6 +67,9 @@ template static void runTestMatchAny_1() { for (int i = 0; i != warpSize; ++i) { REQUIRE(compareMaskEqual(Output, Expected, i, warpSize)); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } template __global__ void matchAny_2(T* Input, unsigned long long* Output) { @@ -124,6 +127,9 @@ template static void runTestMatchAny_2() { for (int i = 0; i != warpSize; ++i) { REQUIRE(compareMaskEqual(Output, Expected, i, warpSize)); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } __global__ void matchAny_3(int* Input, int* Output) { @@ -172,6 +178,9 @@ static void runTestMatchAny_3() { for (size_t i = 0; i < Output.size(); i++) { REQUIRE(Output[i] == Expected[i]); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } /** diff --git a/projects/hip-tests/catch/unit/warp/hipShflSyncDownTests.cc b/projects/hip-tests/catch/unit/warp/hipShflSyncDownTests.cc index f0258a39a5..1868aefe62 100644 --- a/projects/hip-tests/catch/unit/warp/hipShflSyncDownTests.cc +++ b/projects/hip-tests/catch/unit/warp/hipShflSyncDownTests.cc @@ -56,6 +56,9 @@ template static void runTestShflDown_1() { for (int i = 0; i != warpSize; ++i) { REQUIRE(compareEqual(Output[i], Expected[i])); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } // Use the mask argument to divide the warp into groups of 12 threads, and then @@ -96,6 +99,9 @@ template static void runTestShflDown_2() { for (int i = 0; i != warpSize; ++i) { REQUIRE(compareEqual(Output[i], Expected[i])); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } template __global__ void shflDown_3(T* Input, T* Output) { @@ -137,6 +143,9 @@ template static void runTestShflDown_3() { for (int i = 0; i != warpSize; ++i) { REQUIRE(compareEqual(Output[i], Expected[i])); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } __global__ void shflDown_4(int* Input, int* Output) { @@ -193,6 +202,9 @@ static void runTestShflDown_4() { for (size_t i = 0; i < Output.size(); i++) { REQUIRE(Output[i] == Expected[i]); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } /** diff --git a/projects/hip-tests/catch/unit/warp/hipShflSyncTests.cc b/projects/hip-tests/catch/unit/warp/hipShflSyncTests.cc index b9b5ee84b0..20d622f7f6 100644 --- a/projects/hip-tests/catch/unit/warp/hipShflSyncTests.cc +++ b/projects/hip-tests/catch/unit/warp/hipShflSyncTests.cc @@ -56,6 +56,9 @@ template static void runTestShfl_1() { for (int i = 0; i != warpSize; ++i) { REQUIRE(compareEqual(Output[i], Expected[i])); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } template __global__ void shfl_2(T* Input, T* Output) { @@ -95,6 +98,9 @@ template static void runTestShfl_2() { for (int i = 0; i != warpSize; ++i) { REQUIRE(compareEqual(Output[i], Expected[i])); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } __global__ void shfl_3(int* Input, int* Output) { @@ -138,6 +144,9 @@ static void runTestShfl_3() { for (size_t i = 0; i < Output.size(); i++) { REQUIRE(Output[i] == Expected[i]); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } /** diff --git a/projects/hip-tests/catch/unit/warp/hipShflSyncUpTests.cc b/projects/hip-tests/catch/unit/warp/hipShflSyncUpTests.cc index 348204cb7c..d55ee500da 100644 --- a/projects/hip-tests/catch/unit/warp/hipShflSyncUpTests.cc +++ b/projects/hip-tests/catch/unit/warp/hipShflSyncUpTests.cc @@ -53,6 +53,9 @@ template static void runTestShflUp_1() { for (int i = 0; i != warpSize; ++i) { REQUIRE(compareEqual(Output[i], Expected[i])); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } template __global__ void shflUp_2(T* Input, T* Output) { @@ -89,6 +92,9 @@ template static void runTestShflUp_2() { for (int i = 0; i != warpSize; ++i) { REQUIRE(compareEqual(Output[i], Expected[i])); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } template __global__ void shflUp_3(T* Input, T* Output) { @@ -131,6 +137,9 @@ template static void runTestShflUp_3() { for (int i = 0; i != warpSize; ++i) { REQUIRE(compareEqual(Output[i], Expected[i])); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } __global__ void shflUp_4(int* Input, int* Output) { @@ -175,6 +184,9 @@ static void runTestShflUp_4() { for (size_t i = 0; i < Output.size(); i++) { REQUIRE(Output[i] == Expected[i]); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } /** diff --git a/projects/hip-tests/catch/unit/warp/hipShflSyncXorTests.cc b/projects/hip-tests/catch/unit/warp/hipShflSyncXorTests.cc index 2d2a1429a1..b70b636f31 100644 --- a/projects/hip-tests/catch/unit/warp/hipShflSyncXorTests.cc +++ b/projects/hip-tests/catch/unit/warp/hipShflSyncXorTests.cc @@ -261,6 +261,9 @@ static void runTestShflXor_4() { for (size_t i = 0; i < Output.size(); i++) { REQUIRE(Output[i] == Expected[i]); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } /** diff --git a/projects/hip-tests/catch/unit/warp/hipVoteSyncTests.cc b/projects/hip-tests/catch/unit/warp/hipVoteSyncTests.cc index 7bbde18f33..0bb0a8e2a5 100644 --- a/projects/hip-tests/catch/unit/warp/hipVoteSyncTests.cc +++ b/projects/hip-tests/catch/unit/warp/hipVoteSyncTests.cc @@ -49,6 +49,9 @@ static void runTestAny_1() { for (int i = 0; i != warpSize; ++i) { REQUIRE(Output[i] == Expected[i]); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } __global__ void any_2(int* Input, int* Output) { @@ -83,6 +86,9 @@ static void runTestAny_2_w64() { for (int i = 0; i != warpSize; ++i) { REQUIRE(Output[i] == Expected[i]); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } static void runTestAny_2_w32() { @@ -112,6 +118,9 @@ static void runTestAny_2_w32() { for (int i = 0; i != warpSize; ++i) { REQUIRE(Output[i] == Expected[i]); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } __global__ void any_3(int* Input, int* Output) { @@ -144,6 +153,9 @@ static void runTestAny_3() { for (int i = 0; i != warpSize; ++i) { REQUIRE(Output[i] == Expected[i]); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } __global__ void any_4(int* Input, int* Output) { @@ -192,6 +204,9 @@ static void runTestAny_4() { for (int i = 0; i != warpSize; ++i) { REQUIRE(Output[i] == Expected[i]); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } __global__ void all_1(int* Input, int* Output) { @@ -225,6 +240,9 @@ static void runTestAll_1_w64() { for (int i = 0; i != warpSize; ++i) { REQUIRE(Output[i] == Expected[i]); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } static void runTestAll_1_w32() { @@ -252,6 +270,9 @@ static void runTestAll_1_w32() { for (int i = 0; i != warpSize; ++i) { REQUIRE(Output[i] == Expected[i]); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } __global__ void all_2(int* Input, int* Output) { @@ -284,6 +305,9 @@ static void runTestAll_2() { for (int i = 0; i != warpSize; ++i) { REQUIRE(Output[i] == Expected[i]); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } __global__ void all_3(int* Input, int* Output) { @@ -316,6 +340,9 @@ static void runTestAll_3() { for (int i = 0; i != warpSize; ++i) { REQUIRE(Output[i] == Expected[i]); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } __global__ void all_4(int* Input, int* Output) { @@ -363,6 +390,9 @@ static void runTestAll_4() { for (int i = 0; i != warpSize; ++i) { REQUIRE(Output[i] == Expected[i]); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } __global__ void ballot_1(int* Input, unsigned long long* Output) { @@ -408,6 +438,9 @@ static void runTestBallot_1() { for (int i = 0; i != warpSize; ++i) { REQUIRE(compareMaskEqual(Output, Expected, i, warpSize)); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } __global__ void ballot_2(int* Input, unsigned long long* Output) { @@ -454,6 +487,9 @@ static void runTestBallot_2() { for (int i = 0; i != warpSize; ++i) { REQUIRE(compareMaskEqual(Output, Expected, i, warpSize)); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } __global__ void ballot_3(int* Input, unsigned long long* Output) { @@ -500,6 +536,9 @@ static void runTestBallot_3() { for (int i = 0; i != warpSize; ++i) { REQUIRE(compareMaskEqual(Output, Expected, i, warpSize)); } + + HIP_CHECK(hipFree(d_Input)); + HIP_CHECK(hipFree(d_Output)); } /**