From 41acbf98be9ae0db675b46065d00950ef3e6de6c Mon Sep 17 00:00:00 2001 From: ROCm CI Service Account <66695075+rocm-ci@users.noreply.github.com> Date: Tue, 1 Nov 2022 06:08:26 +0530 Subject: [PATCH 1/2] SWDEV-360437 - fix tests with hipGetLastError() (#3040) Unit_hipFreeImplicitSyncDev* Unit_hipFreeImplicitSyncHost* Unit_hipFreeImplicitSyncArray* Unit_hipStreamCreateWithFlags_DefaultStreamInteraction Unit_hipStreamSynchronize_NullStreamAndStreamPerThread Change-Id: I382cf13bfe15bd8b1af657bdbb8c4f9043ea6f17 --- tests/catch/include/hip_test_common.hh | 1 - tests/catch/unit/event/hipEventSynchronize.cc | 4 ++-- 2 files changed, 2 insertions(+), 3 deletions(-) diff --git a/tests/catch/include/hip_test_common.hh b/tests/catch/include/hip_test_common.hh index ed8412b673..5f7b197596 100644 --- a/tests/catch/include/hip_test_common.hh +++ b/tests/catch/include/hip_test_common.hh @@ -397,7 +397,6 @@ static inline void runKernelForDuration(std::chrono::milliseconds duration, static size_t ticksPerSecond = findTicksPerSecond(); const auto millis = duration.count(); hipLaunchKernelGGL(waitKernel, dim3(1), dim3(1), 0, stream, ticksPerSecond * millis / 1000); - HIP_CHECK(hipGetLastError()); } } // namespace HipTest diff --git a/tests/catch/unit/event/hipEventSynchronize.cc b/tests/catch/unit/event/hipEventSynchronize.cc index e3c07bb395..a347badb22 100644 --- a/tests/catch/unit/event/hipEventSynchronize.cc +++ b/tests/catch/unit/event/hipEventSynchronize.cc @@ -110,13 +110,13 @@ TEST_CASE("Unit_hipEventSynchronize_NoEventRecord_Positive") { // Record the end_event HIP_CHECK(hipEventRecord(end_event, NULL)); + // End event has not been completed + HIP_CHECK_ERROR(hipEventQuery(end_event), hipErrorNotReady); // When hipEventSynchronized is called on event that has not been recorded, // the function returns immediately HIP_CHECK(hipEventSynchronize(dummy_event)); - // End event has not been completed - HIP_CHECK_ERROR(hipEventQuery(end_event), hipErrorNotReady); // Wait for end_event to complete HIP_CHECK(hipEventSynchronize(end_event)); From b0d68926920a6ab0c936ce6ed46728ed5ad21124 Mon Sep 17 00:00:00 2001 From: ROCm CI Service Account <66695075+rocm-ci@users.noreply.github.com> Date: Tue, 1 Nov 2022 10:38:38 +0530 Subject: [PATCH 2/2] SWDEV-347670 - Enable GPU StreamWrite and StreamWait tests on Windows (#3031) This patch skips Unit_hipStreamValue_*_Blocking test cases because of a soft hang Stream write and stream wait passing tests can be enabled in the device attribute. Unit_hipStreamValue_*_Blocking tests will be enabled when the hang is fixed Change-Id: Ieea831162208c4ec172b35a0421c66f6b1a21ed7 --- .../config/config_amd_windows_MI2xx.json | 19 ++++++++++++++++++- .../config/config_amd_windows_common.json | 19 ++++++++++++++++++- tests/catch/unit/event/Unit_hipEventIpc.cc | 6 +++++- 3 files changed, 41 insertions(+), 3 deletions(-) diff --git a/tests/catch/hipTestMain/config/config_amd_windows_MI2xx.json b/tests/catch/hipTestMain/config/config_amd_windows_MI2xx.json index e6b4364a87..013f7c67d8 100644 --- a/tests/catch/hipTestMain/config/config_amd_windows_MI2xx.json +++ b/tests/catch/hipTestMain/config/config_amd_windows_MI2xx.json @@ -71,6 +71,23 @@ "Unit_hipStreamSynchronize_FinishWork", "Unit_hipStreamSynchronize_NullStreamAndStreamPerThread", "Unit_hipMultiThreadDevice_NearZero", - "Unit_hipStreamPerThread_DeviceReset_1" + "Unit_hipStreamPerThread_DeviceReset_1", + "Unit_hipStreamValue_Wait32_Blocking_Mask_Gte", + "Unit_hipStreamValue_Wait32_Blocking_Mask_Eq_1", + "Unit_hipStreamValue_Wait32_Blocking_Mask_Eq_2", + "Unit_hipStreamValue_Wait32_Blocking_Mask_And", + "Unit_hipStreamValue_Wait32_Blocking_NoMask_Eq", + "Unit_hipStreamValue_Wait32_Blocking_NoMask_Gte", + "Unit_hipStreamValue_Wait32_Blocking_NoMask_And", + "Unit_hipStreamValue_Wait32_Blocking_NoMask_Nor", + "Unit_hipStreamValue_Wait64_Blocking_Mask_Gte_1", + "Unit_hipStreamValue_Wait64_Blocking_Mask_Gte_2", + "Unit_hipStreamValue_Wait64_Blocking_Mask_Eq_1", + "Unit_hipStreamValue_Wait64_Blocking_Mask_Eq_2", + "Unit_hipStreamValue_Wait64_Blocking_Mask_And", + "Unit_hipStreamValue_Wait64_Blocking_NoMask_Gte", + "Unit_hipStreamValue_Wait64_Blocking_NoMask_Eq", + "Unit_hipStreamValue_Wait64_Blocking_NoMask_And", + "Unit_hipStreamValue_Wait64_Blocking_NoMask_Nor" ] } diff --git a/tests/catch/hipTestMain/config/config_amd_windows_common.json b/tests/catch/hipTestMain/config/config_amd_windows_common.json index 002666b345..07a4714117 100644 --- a/tests/catch/hipTestMain/config/config_amd_windows_common.json +++ b/tests/catch/hipTestMain/config/config_amd_windows_common.json @@ -78,6 +78,23 @@ "Unit_hipGraphMemcpyNodeSetParamsToSymbol_Functional", "Unit_hipStreamQuery_WithPendingWork", "Unit_hipStreamWaitEvent_DifferentStreams", - "Unit_hipStreamQuery_WithFinishedWork" + "Unit_hipStreamQuery_WithFinishedWork", + "Unit_hipStreamValue_Wait32_Blocking_Mask_Gte", + "Unit_hipStreamValue_Wait32_Blocking_Mask_Eq_1", + "Unit_hipStreamValue_Wait32_Blocking_Mask_Eq_2", + "Unit_hipStreamValue_Wait32_Blocking_Mask_And", + "Unit_hipStreamValue_Wait32_Blocking_NoMask_Eq", + "Unit_hipStreamValue_Wait32_Blocking_NoMask_Gte", + "Unit_hipStreamValue_Wait32_Blocking_NoMask_And", + "Unit_hipStreamValue_Wait32_Blocking_NoMask_Nor", + "Unit_hipStreamValue_Wait64_Blocking_Mask_Gte_1", + "Unit_hipStreamValue_Wait64_Blocking_Mask_Gte_2", + "Unit_hipStreamValue_Wait64_Blocking_Mask_Eq_1", + "Unit_hipStreamValue_Wait64_Blocking_Mask_Eq_2", + "Unit_hipStreamValue_Wait64_Blocking_Mask_And", + "Unit_hipStreamValue_Wait64_Blocking_NoMask_Gte", + "Unit_hipStreamValue_Wait64_Blocking_NoMask_Eq", + "Unit_hipStreamValue_Wait64_Blocking_NoMask_And", + "Unit_hipStreamValue_Wait64_Blocking_NoMask_Nor" ] } diff --git a/tests/catch/unit/event/Unit_hipEventIpc.cc b/tests/catch/unit/event/Unit_hipEventIpc.cc index cc0d70ae71..1081327f8c 100644 --- a/tests/catch/unit/event/Unit_hipEventIpc.cc +++ b/tests/catch/unit/event/Unit_hipEventIpc.cc @@ -91,9 +91,13 @@ TEST_CASE("Unit_hipEventIpc") { hipEvent_t ipc_event; hipError_t err = hipIpcOpenEventHandle(&ipc_event, ipc_handle); + #if HT_WIN + // always different process Id on Windows + HIP_CHECK(err); + #else // hipIpcOpenEventHandle() should be called in a different process, hence it should fail here REQUIRE(err == hipErrorInvalidContext); - + #endif HIP_CHECK(hipEventDestroy(start)); HIP_CHECK(hipEventDestroy(stop));