From 4e659f42de896faaefda7f635d70bc1791968275 Mon Sep 17 00:00:00 2001 From: SrinivasaRao Date: Fri, 26 Jul 2024 20:45:14 +0530 Subject: [PATCH] SWDEV-466663-[catch2][dtest] _spt api names missing from code coverage-Phase-1 Change-Id: Ifd374fcc242767ab7a9e999dba79f2f3c2ef7c4a --- catch/hipTestMain/config/config_amd_windows | 7 + catch/unit/stream/CMakeLists.txt | 6 +- catch/unit/stream/hipStreamGetFlags_spt.cc | 111 +++++++++++++ catch/unit/stream/hipStreamGetPriority_spt.cc | 156 ++++++++++++++++++ catch/unit/stream/hipStreamQuery_spt.cc | 151 +++++++++++++++++ catch/unit/stream/hipStreamSynchronize_spt.cc | 139 ++++++++++++++++ 6 files changed, 569 insertions(+), 1 deletion(-) create mode 100644 catch/unit/stream/hipStreamGetFlags_spt.cc create mode 100644 catch/unit/stream/hipStreamGetPriority_spt.cc create mode 100644 catch/unit/stream/hipStreamQuery_spt.cc create mode 100644 catch/unit/stream/hipStreamSynchronize_spt.cc diff --git a/catch/hipTestMain/config/config_amd_windows b/catch/hipTestMain/config/config_amd_windows index 5746e17e3c..7ea53395ab 100644 --- a/catch/hipTestMain/config/config_amd_windows +++ b/catch/hipTestMain/config/config_amd_windows @@ -1174,6 +1174,13 @@ "Unit_hipHostAlloc_Negative_NonCoherent", "Unit_hipHostAlloc_Negative_Coherent", "Unit_hipHostAlloc_Negative_NumaUser", + "=== Following tests disabled due to SWDEV-486363", + "Unit_hipStreamQuery_spt_WithFinishedWork", + "Unit_hipStreamQuery_spt_NegativeCases", + "Unit_hipStreamQuery_spt_WithPendingWork", + "Unit_hipStreamSynchronize_spt_FinishWork", + "Unit_hipStreamSynchronize_spt_SynchronizeStreamAndQueryNullStream", + "====================================================", #endif "End of json" ] diff --git a/catch/unit/stream/CMakeLists.txt b/catch/unit/stream/CMakeLists.txt index 0e98be542c..08807754b3 100644 --- a/catch/unit/stream/CMakeLists.txt +++ b/catch/unit/stream/CMakeLists.txt @@ -24,7 +24,11 @@ set(TEST_SRC if(HIP_PLATFORM MATCHES "amd") set(TEST_SRC ${TEST_SRC} hipStreamGetCUMask.cc hipStreamWithCUMask.cc - hipStreamACb_MultiThread.cc hipStreamWaitEvent.cc) + hipStreamACb_MultiThread.cc hipStreamWaitEvent.cc + hipStreamGetFlags_spt.cc + hipStreamGetPriority_spt.cc + hipStreamQuery_spt.cc + hipStreamSynchronize_spt.cc) else() set(TEST_SRC ${TEST_SRC} diff --git a/catch/unit/stream/hipStreamGetFlags_spt.cc b/catch/unit/stream/hipStreamGetFlags_spt.cc new file mode 100644 index 0000000000..0b7cd30459 --- /dev/null +++ b/catch/unit/stream/hipStreamGetFlags_spt.cc @@ -0,0 +1,111 @@ +/* +Copyright (c) 2024 Advanced Micro Devices, Inc. All rights reserved. +Permission is hereby granted, free of charge, to any person obtaining a copy +of this software and associated documentation files (the "Software"), to deal +in the Software without restriction, including without limitation the rights +to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +copies of the Software, and to permit persons to whom the Software is +furnished to do so, subject to the following conditions: +The above copyright notice and this permission notice shall be included in +all copies or substantial portions of the Software. +THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +THE SOFTWARE. +*/ +#include +#include +/** + * @addtogroup hipStreamGetFlags_spt hipStreamGetFlags_spt + * @{ + * @ingroup StreamTest + * `hipStreamGetFlags_spt(hipStream_t stream, unsigned int* flags)` - + * Return flags associated with a stream on stream per thread + */ + + /** + * Test Description + * ------------------------ + * - Basic test to verify that hipStreamGetFlags_spt returns the same + * flags that were used to create the stream. + * Test source + * ------------------------ + * - /unit/stream/hipStreamGetFlags_spt.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.2 + */ +TEST_CASE("Unit_hipStreamGetFlags_spt_Basic") { + unsigned int expectedFlag = GENERATE(hipStreamDefault, hipStreamNonBlocking); + unsigned int returnedFlags; + hipStream_t stream; + + HIP_CHECK(hipStreamCreateWithFlags(&stream, expectedFlag)); + HIP_CHECK(hipStreamGetFlags_spt(stream, &returnedFlags)); + REQUIRE((returnedFlags & expectedFlag) == expectedFlag); + HIP_CHECK(hipStreamDestroy(stream)); +} + /** + * Test Description + * ------------------------ + * - Tests to verify Negative scenarios for the hipStreamGetFlags_spt API + * Test source + * ------------------------ + * - /unit/stream/hipStreamGetFlags_spt.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.2 + */ + +TEST_CASE("Unit_hipStreamGetFlags_spt_Negative") { + hipStream_t validStream; + unsigned int flags; + + HIP_CHECK(hipStreamCreate(&validStream)); + + SECTION("Nullptr Stream && Valid Flags") { + #if HT_AMD + HIP_CHECK_ERROR(hipStreamGetFlags_spt(nullptr, &flags), hipSuccess); + #elif HT_NVIDIA + HIP_CHECK(hipStreamGetFlags_spt(nullptr, &flags)); + #endif + } + + SECTION("Valid Stream && Nullptr Flags") { + HIP_CHECK_ERROR(hipStreamGetFlags_spt(validStream, nullptr), + hipErrorInvalidValue); + } + + HIP_CHECK(hipStreamDestroy(validStream)); +} + /** + * Test Description + * ------------------------ + * - Test flag value when streams created with CUMask + * Test source + * ------------------------ + * - /unit/stream/hipStreamGetFlags_spt.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.2 + */ + +#if HT_AMD +TEST_CASE("Unit_hipStreamGetFlags_spt_StreamsCreatedWithCUMask") { + hipStream_t stream; + unsigned int flags; + const uint32_t cuMask = 0xffffffff; + HIP_CHECK(hipExtStreamCreateWithCUMask(&stream, 1, &cuMask)); + HIP_CHECK(hipStreamGetFlags_spt(stream, &flags)); + REQUIRE(flags == hipStreamDefault); + HIP_CHECK(hipStreamDestroy(stream)); +} +#endif + +/** +* End doxygen group StreamTest. +* @} +*/ diff --git a/catch/unit/stream/hipStreamGetPriority_spt.cc b/catch/unit/stream/hipStreamGetPriority_spt.cc new file mode 100644 index 0000000000..b58c32ec17 --- /dev/null +++ b/catch/unit/stream/hipStreamGetPriority_spt.cc @@ -0,0 +1,156 @@ +/* +Copyright (c) 2024 Advanced Micro Devices, Inc. All rights reserved. +Permission is hereby granted, free of charge, to any person obtaining a copy +of this software and associated documentation files (the "Software"), to deal +in the Software without restriction, including without limitation the rights +to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +copies of the Software, and to permit persons to whom the Software is +furnished to do so, subject to the following conditions: +The above copyright notice and this permission notice shall be included in +all copies or substantial portions of the Software. +THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +THE SOFTWARE. +*/ + +#include +#include + +/** + * @addtogroup hipStreamGetPriority_spt hipStreamGetPriority_spt + * @{ + * @ingroup StreamTest + * `hipStreamGetPriority_spt(hipStream_t stream, int* priority)` - + * Query the priority of a stream + */ + + /** + * Test Description + * ------------------------ + * - Test to create a stream and check priority. + * Test source + * ------------------------ + * - /unit/stream/hipStreamGetPriority_spt.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.2 + */ +TEST_CASE("Unit_hipStreamGetPriority_spt_BasicTst") { + int priority_low = 0; + int priority_high = 0; + int devID = GENERATE(range(0, HipTest::getDeviceCount())); + HIP_CHECK(hipSetDevice(devID)); + HIP_CHECK(hipDeviceGetStreamPriorityRange(&priority_low, &priority_high)); + hipStream_t stream{}; + int priority = 0; + SECTION("Null Stream") { + HIP_CHECK(hipStreamGetPriority_spt(nullptr, &priority)); + // valid priority + REQUIRE(priority_low >= priority); + REQUIRE(priority >= priority_high); + } + SECTION("Created Stream") { + SECTION("Default Priority") { + HIP_CHECK(hipStreamCreate(&stream)); + HIP_CHECK(hipStreamGetPriority_spt(stream, &priority)); + // valid priority + // Lower the value higher the priority, + // higher the value lower the priority + REQUIRE(priority_low >= priority); + REQUIRE(priority >= priority_high); + } + SECTION("High Priority") { + HIP_CHECK(hipStreamCreateWithPriority(&stream, hipStreamDefault, + priority_high)); + HIP_CHECK(hipStreamGetPriority_spt(stream, &priority)); + REQUIRE(priority == priority_high); + } + SECTION("Higher Priority") { + HIP_CHECK(hipStreamCreateWithPriority(&stream, hipStreamNonBlocking, + priority_high - 1)); + HIP_CHECK(hipStreamGetPriority_spt(stream, &priority)); + REQUIRE(priority == priority_high); + } + SECTION("Low Priority") { + HIP_CHECK(hipStreamCreateWithPriority(&stream, hipStreamDefault, + priority_low)); + HIP_CHECK(hipStreamGetPriority_spt(stream, &priority)); + REQUIRE(priority_low == priority); + } + SECTION("Lower Priority") { + HIP_CHECK(hipStreamCreateWithPriority(&stream, hipStreamNonBlocking, + priority_low + 1)); + HIP_CHECK(hipStreamGetPriority_spt(stream, &priority)); + REQUIRE(priority_low == priority); + } + HIP_CHECK(hipStreamDestroy(stream)); + } +} + + /** + * Test Description + * ------------------------ + * - Test check stream priority with negative cases. + * Test source + * ------------------------ + * - /unit/stream/hipStreamGetPriority_spt.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.2 + */ +TEST_CASE("Unit_hipStreamGetPriority_spt_NegativeCases") { + // stream as nullptr and priority as nullptr + auto res = hipStreamGetPriority_spt(nullptr, nullptr); + REQUIRE(res == hipErrorInvalidValue); + // priority as nullptr + hipStream_t stream = nullptr; + HIP_CHECK(hipStreamCreate(&stream)); + res = hipStreamGetPriority_spt(stream, nullptr); + REQUIRE(res == hipErrorInvalidValue); + HIP_CHECK(hipStreamDestroy(stream)); + // stream as nullptr + int priority = -1; + HIP_CHECK(hipStreamGetPriority_spt(nullptr, &priority)); +} + /** + * Test Description + * ------------------------ + * - Create stream with CUMask and check priority is returned as expected. + * Test source + * ------------------------ + * - /unit/stream/hipStreamGetPriority_spt.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.2 + */ + +#if HT_AMD +TEST_CASE("Unit_hipStreamGetPriority_spt_StreamsWithCUMask") { + hipStream_t stream{}; + int priority = 0; + int priority_normal = 0; + int priority_low = 0; + int priority_high = 0; + // Test is to get the Stream Priority Range + HIP_CHECK(hipDeviceGetStreamPriorityRange(&priority_low, &priority_high)); + priority_normal = (priority_low + priority_high) / 2; + // Check if priorities are indeed supported + REQUIRE_FALSE(priority_low == priority_high); + // Creating a stream with hipExtStreamCreateWithCUMask and checking + // priority. + const uint32_t cuMask = 0xffffffff; + HIP_CHECK(hipExtStreamCreateWithCUMask(&stream, 1, &cuMask)); + HIP_CHECK(hipStreamGetPriority_spt(stream, &priority)); + REQUIRE_FALSE(priority_normal != priority); + HIP_CHECK(hipStreamDestroy(stream)); +} +#endif + +/** +* End doxygen group StreamTest. +* @} +*/ diff --git a/catch/unit/stream/hipStreamQuery_spt.cc b/catch/unit/stream/hipStreamQuery_spt.cc new file mode 100644 index 0000000000..37b011b62f --- /dev/null +++ b/catch/unit/stream/hipStreamQuery_spt.cc @@ -0,0 +1,151 @@ +/* +Copyright (c) 2024 Advanced Micro Devices, Inc. All rights reserved. +Permission is hereby granted, free of charge, to any person obtaining a copy +of this software and associated documentation files (the "Software"), to deal +in the Software without restriction, including without limitation the rights +to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +copies of the Software, and to permit persons to whom the Software is +furnished to do so, subject to the following conditions: +The above copyright notice and this permission notice shall be included in +all copies or substantial portions of the Software. +THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +THE SOFTWARE. +*/ +#include +#include +#include "streamCommon.hh" +/** + * @addtogroup hipStreamQuery_spt hipStreamQuery_spt + * @{ + * @ingroup StreamTest + * `hipStreamQuery_spt(hipStream_t stream)` - + * Returns status of a stream + */ + +/** + * Test Description + * ------------------------ + * - Test to query a stream with no work returns hipSuccess or not. + * Test source + * ------------------------ + * - /unit/stream/hipStreamQuery_spt.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.2 + */ +TEST_CASE("Unit_hipStreamQuery_spt_WithNoWork") { + hipStream_t stream{nullptr}; + + SECTION("Null Stream") { + HIP_CHECK(hipStreamQuery_spt(stream)); + } + + SECTION("Created Stream") { + HIP_CHECK(hipStreamCreate(&stream)); + HIP_CHECK(hipStreamQuery_spt(stream)); + HIP_CHECK(hipStreamDestroy(stream)); + } +} +/** + * Test Description + * ------------------------ + * - Test to query a stream with finished work returns hipSuccess or not. + * Test source + * ------------------------ + * - /unit/stream/hipStreamQuery_spt.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.2 + */ +TEST_CASE("Unit_hipStreamQuery_spt_WithFinishedWork") { + hipStream_t stream{nullptr}; + + SECTION("Null Stream") { + hip::stream::empty_kernel<<>>(); + HIP_CHECK(hipStreamSynchronize(stream)); + + HIP_CHECK(hipStreamQuery_spt(stream)); + } + + SECTION("Created Stream") { + HIP_CHECK(hipStreamCreate(&stream)); + hip::stream::empty_kernel<<>>(); + HIP_CHECK(hipStreamSynchronize(stream)); + + HIP_CHECK(hipStreamQuery_spt(stream)); + HIP_CHECK(hipStreamDestroy(stream)); + } +} +/** + * Test Description + * ------------------------ + * - Negative cases to query a stream. + * Test source + * ------------------------ + * - /unit/stream/hipStreamQuery_spt.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.2 + */ +#if HT_AMD +TEST_CASE("Unit_hipStreamQuery_spt_NegativeCases") { + SECTION("Query Destroyed Stream") { + hipStream_t stream{nullptr}; + HIP_CHECK(hipStreamCreate(&stream)); + HIP_CHECK(hipStreamDestroy(stream)); + HIP_CHECK_ERROR(hipStreamQuery_spt(stream), hipErrorContextIsDestroyed); + } + SECTION("Query Uninitialized Stream") { + hipStream_t streamUnInit{reinterpret_cast(0xFFFF)}; + HIP_CHECK_ERROR(hipStreamQuery_spt(streamUnInit), + hipErrorContextIsDestroyed); + } + SECTION("Submit Work On Stream And Query Null Stream") { + hipStream_t ValidStream; + HIP_CHECK(hipStreamCreate(&ValidStream)); + + HIP_CHECK(hipStreamQuery_spt(hip::nullStream)); + LaunchDelayKernel(std::chrono::milliseconds(500), ValidStream); + HIP_CHECK_ERROR(hipStreamQuery_spt(hip::nullStream), hipSuccess); + HIP_CHECK(hipDeviceSynchronize()); + HIP_CHECK(hipStreamDestroy(ValidStream)); + } + SECTION("Submit work to the nullStream") { + HIP_CHECK(hipStreamQuery_spt(hip::nullStream)); + LaunchDelayKernel(std::chrono::milliseconds(500), hip::nullStream); + HIP_CHECK_ERROR(hipStreamQuery_spt(hip::nullStream), hipSuccess); + HIP_CHECK(hipStreamSynchronize(hip::nullStream)); + } +} +/** + * Test Description + * ------------------------ + * - Test to querying a stream with pending work returns hipErrorNotReady or not. + * Test source + * ------------------------ + * - /unit/stream/hipStreamQuery_spt.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.2 + */ +TEST_CASE("Unit_hipStreamQuery_spt_WithPendingWork") { + hipStream_t waitingStream{nullptr}; + HIP_CHECK(hipStreamCreate(&waitingStream)); + LaunchDelayKernel(std::chrono::milliseconds(500), waitingStream); + HIP_CHECK_ERROR(hipStreamQuery_spt(waitingStream), hipErrorNotReady); + HIP_CHECK(hipStreamSynchronize(waitingStream)); + HIP_CHECK(hipStreamQuery_spt(waitingStream)); + HIP_CHECK(hipStreamDestroy(waitingStream)); +} +#endif + +/** +* End doxygen group StreamTest. +* @} +*/ + diff --git a/catch/unit/stream/hipStreamSynchronize_spt.cc b/catch/unit/stream/hipStreamSynchronize_spt.cc new file mode 100644 index 0000000000..8b97b4e0c4 --- /dev/null +++ b/catch/unit/stream/hipStreamSynchronize_spt.cc @@ -0,0 +1,139 @@ +/* +Copyright (c) 2024 Advanced Micro Devices, Inc. All rights reserved. +Permission is hereby granted, free of charge, to any person obtaining a copy +of this software and associated documentation files (the "Software"), to deal +in the Software without restriction, including without limitation the rights +to use, copy, modify, merge, publish, distribute, sublicense, and/or sell +copies of the Software, and to permit persons to whom the Software is +furnished to do so, subject to the following conditions: +The above copyright notice and this permission notice shall be included in +all copies or substantial portions of the Software. +THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR +IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, +FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE +AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER +LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, +OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN +THE SOFTWARE. +*/ +#include +#include +#include +#include "streamCommon.hh" +/** + * @addtogroup hipStreamSynchronize_spt hipStreamSynchronize_spt + * @{ + * @ingroup StreamTest + * `hipStreamSynchronize_spt(hipStream_t stream)` - + * Query the priority of a stream + */ + + /** + * Test Description + * ------------------------ + * - Test to verify that hipStreamSynchronize_spt handles empty streams properly. + * Test source + * ------------------------ + * - /unit/stream/hipStreamSynchronize_spt.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.2 + */ +TEST_CASE("Unit_hipStreamSynchronize_spt_EmptyStream") { + hipStream_t stream; + HIP_CHECK(hipStreamCreate(&stream)); + HIP_CHECK(hipStreamSynchronize_spt(stream)); + HIP_CHECK(hipStreamDestroy(stream)); +} + /** + * Test Description + * ------------------------ + * - Check that synchronization of uninitialized stream sets its status to + * hipErrorContextIsDestroyed + * Test source + * ------------------------ + * - /unit/stream/hipStreamSynchronize_spt.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.2 + */ + +#if HT_AMD +TEST_CASE("Unit_hipStreamSynchronize_spt_UninitializedStream") { + hipStream_t stream{reinterpret_cast(0xFFFF)}; + HIP_CHECK_ERROR(hipStreamSynchronize_spt(stream), hipErrorContextIsDestroyed); +} + /** + * Test Description + * ------------------------ + * - Check that all work executing in a stream is finished after a call to + * hipStreamSynchronize_spt. + * Test source + * ------------------------ + * - /unit/stream/hipStreamSynchronize_spt.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.2 + */ +TEST_CASE("Unit_hipStreamSynchronize_spt_FinishWork") { + hipStream_t explicitStream = reinterpret_cast(-1); + hipStream_t stream = GENERATE_COPY(explicitStream, hip::nullStream, + hip::streamPerThread); + if (explicitStream) { + HIP_CHECK(hipStreamCreate(&stream)); + } + + LaunchDelayKernel(std::chrono::milliseconds(500), stream); + HIP_CHECK(hipStreamSynchronize_spt(stream)); + HIP_CHECK(hipStreamQuery(stream)); + + if (explicitStream) { + HIP_CHECK(hipStreamDestroy(stream)); + } +} + /** + * Test Description + * ------------------------ + * - Check that synchronizing one stream does implicitly synchronize + * other streams. Check that submiting work to the nullStream does + * not affect synchronization of other streams. Check that querying + * the nullStream does not affect synchronization of other streams. + * Test source + * ------------------------ + * - /unit/stream/hipStreamSynchronize_spt.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.2 + */ + +TEST_CASE("Unit_hipStreamSynchronize_spt_SynchronizeStreamAndQueryNullStream") { + hipStream_t stream1; + hipStream_t stream2; + + HIP_CHECK(hipStreamCreate(&stream1)); + HIP_CHECK(hipStreamCreate(&stream2)); + + LaunchDelayKernel(std::chrono::milliseconds(500), stream1); + LaunchDelayKernel(std::chrono::milliseconds(2000), stream2); + hip::stream::empty_kernel<<<1, 1, 0, hip::nullStream>>>(); + + HIP_CHECK_ERROR(hipStreamQuery(stream1), hipErrorNotReady); + HIP_CHECK_ERROR(hipStreamQuery(stream2), hipErrorNotReady); + HIP_CHECK_ERROR(hipStreamQuery(hip::nullStream), hipErrorNotReady); + + HIP_CHECK(hipStreamSynchronize_spt(stream1)); + HIP_CHECK(hipStreamQuery(stream1)); + HIP_CHECK_ERROR(hipStreamQuery(stream2), hipErrorNotReady); + HIP_CHECK_ERROR(hipStreamQuery(hip::nullStream), hipErrorNotReady); + + HIP_CHECK(hipStreamSynchronize_spt(stream2)); + HIP_CHECK(hipStreamQuery(stream2)); + + HIP_CHECK(hipStreamDestroy(stream1)); + HIP_CHECK(hipStreamDestroy(stream2)); +} +#endif +/** +* End doxygen group StreamTest. +* @} +*/