diff --git a/catch/hipTestMain/config/config_amd_linux b/catch/hipTestMain/config/config_amd_linux index 398ced33b3..75d3ffb608 100644 --- a/catch/hipTestMain/config/config_amd_linux +++ b/catch/hipTestMain/config/config_amd_linux @@ -132,14 +132,6 @@ "Unit_hipEventIpc", "=== SWDEV-427101:Below test fails randomly in PSDB ===", "Unit_deviceAllocation_InOneThread_AccessInAllThreads", - "=== StreamCreate Performance tests may potentially fail on all configs ===", - "Unit_hipStreamCreate_Performance", - "Unit_hipStreamCreate_WithFlagsPerformance_Nonblocking", - "Unit_hipStreamCreate_WithFlagsPerformance_Default", - "Unit_hipStreamCreate_WithPriorityPerformance_Nonblocking_low", - "Unit_hipStreamCreate_WithPriorityPerformance_Nonblocking_high", - "Unit_hipStreamCreate_WithPriorityPerformance_Default_low", - "Unit_hipStreamCreate_WithPriorityPerformance_Default_high", "=== Patch which removes the typetraits implementation from std namespace in hiprtc is reverted ===", "Unit_hiprtc_stdheaders", #endif @@ -221,12 +213,6 @@ "=== SWDEV-432556:Below test failed in stress test on 10/11/23 ===", "Unit_hipDeviceGetUuid_From_RocmInfo", #endif - #if defined NAVI21 - "=== Below tests failed in stress test on 11/10/23 ===", - "Unit_hipStreamCreate_Performance", - "Unit_hipStreamCreate_WithFlagsPerformance_Nonblocking", - "Unit_hipStreamCreate_WithFlagsPerformance_Default", - #endif #if defined NAVI3X "=== Below tests soft hang in stress test on 13/09/23 ===", "Unit_hipMemsetFunctional_ZeroValue_hipMemsetD16", @@ -238,8 +224,6 @@ "Grid_Group_Getters_Via_Non_Member_Functions_Positive_Basic", "Grid_Group_Sync_Positive_Basic", "dynamic_loading_device_kernels_from_library", - "=== Below tests failed in stress test on 11/10/23 ===", - "Unit_hipStreamCreate_WithPriorityPerformance_Nonblocking_low", #endif "End of json" ] diff --git a/catch/hipTestMain/config/config_amd_windows b/catch/hipTestMain/config/config_amd_windows index cc41c48b0d..ad4995ec99 100644 --- a/catch/hipTestMain/config/config_amd_windows +++ b/catch/hipTestMain/config/config_amd_windows @@ -213,14 +213,6 @@ "Unit_hipGetChannelDesc_Negative_Parameters", "Unit_hipTextureMipmapRef2D_Positive_Check", "Unit_hipTextureMipmapRef2D_Negative_Parameters", - "=== StreamCreate Performance tests may potentially fail on all configs ===", - "Unit_hipStreamCreate_Performance", - "Unit_hipStreamCreate_WithFlagsPerformance_Nonblocking", - "Unit_hipStreamCreate_WithFlagsPerformance_Default", - "Unit_hipStreamCreate_WithPriorityPerformance_Nonblocking_low", - "Unit_hipStreamCreate_WithPriorityPerformance_Nonblocking_high", - "Unit_hipStreamCreate_WithPriorityPerformance_Default_low", - "Unit_hipStreamCreate_WithPriorityPerformance_Default_high", "=== SWDEV-430116:Below tests failed in stress test on 27/10/23 ===", "Unit_hipFreeAsync_negative", "Unit_hipLaunchHostFunc_multistreams", diff --git a/catch/hipTestMain/config/config_nvidia_linux.json b/catch/hipTestMain/config/config_nvidia_linux.json index 6ec6c657ef..ae93da67d6 100644 --- a/catch/hipTestMain/config/config_nvidia_linux.json +++ b/catch/hipTestMain/config/config_nvidia_linux.json @@ -20,11 +20,6 @@ "Unit_ChannelDescriptor_Positive_Basic_4D - long4", "=== Below test fails in external CI for PR https://github.com/ROCm-Developer-Tools/hip-tests/pull/38 ===", "Unit_hipFreeAsync_negative", - "=== Below test fails in external CI for PR https://github.com/ROCm-Developer-Tools/hip-tests/pull/222 ===", - "Unit_hipStreamCreate_Performance", - "Unit_hipStreamCreate_WithPriorityPerformance_Nonblocking_high", - "Unit_hipStreamCreate_WithPriorityPerformance_Default_low", - "Unit_hipStreamCreate_WithPriorityPerformance_Nonblocking_low", "=== Below test fails in external CI for PR https://github.com/ROCm-Developer-Tools/hip-tests/pull/92 ===", "Unit_hipGetTexObjectResourceDesc_positive", "Unit_hipGetTexObjectResourceDesc_Negative_Parameters", @@ -32,7 +27,6 @@ "Unit_hipGetTexObjectTextureDesc_Negative_Parameters", "Unit_hipTexObjectDestroy_positive", "=== Below tests tests fail randomly in PSDB ===", - "Unit_hipStreamCreate_WithPriorityPerformance_Default_high", "Unit_hipGraphInstantiateWithFlags_DependencyGraphDeviceCtxtChg", "Unit_hipGraphUpload_Functional_multidevice_test", "Unit_hipMemcpyParam2D_multiDevice-D2D - char", @@ -43,8 +37,6 @@ "Unit_hipMemcpyParam2DAsync_multiDevice-StreamOnDiffDevice - char", "Unit_hipMemsetFunctional_ZeroValue_hipMemsetD16", "Unit_hipIpcMemAccess_Semaphores", - "Unit_hipStreamCreate_WithFlagsPerformance_Nonblocking", - "Unit_hipStreamCreate_WithFlagsPerformance_Default", "Unit_hipStreamAttachMemAsync_Negative_Parameters", "hipStreamPerThread_CoopLaunch", "hipCGMultiGridGroupType", diff --git a/catch/unit/stream/CMakeLists.txt b/catch/unit/stream/CMakeLists.txt index 5c0f349e27..267b00d38a 100644 --- a/catch/unit/stream/CMakeLists.txt +++ b/catch/unit/stream/CMakeLists.txt @@ -17,7 +17,6 @@ set(TEST_SRC hipStreamACb_StrmSyncTiming.cc hipLaunchHostFunc.cc hipStreamGetDevice.cc - hipStreamCreatePerformance.cc ) if(HIP_PLATFORM MATCHES "amd") diff --git a/catch/unit/stream/hipStreamCreatePerformance.cc b/catch/unit/stream/hipStreamCreatePerformance.cc deleted file mode 100644 index 999a4fe7bf..0000000000 --- a/catch/unit/stream/hipStreamCreatePerformance.cc +++ /dev/null @@ -1,544 +0,0 @@ -/* -Copyright (c) 2023 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 - -/** - * @addtogroup hipStreamCreate hipStreamCreate - * @{ - * @ingroup StreamTest - * `hipError_t hipStreamCreate(hipStream_t* stream)` - - * Create a new asynchronous stream. - */ - -/** - * Test Description - * ------------------------ - * - Test case to verify hipStreamCreate performance by recording below sets of time. - * create 4 set of streams and record that time taken as b1 - * create another 4 set of streams before destroying earlier streams and record time taken as b2 - * destroy streams and then create 4 streams again and record time taken as b3 - * and verify if the condition b1 > b2 > b3 is satisfied. - - * Test source - * ------------------------ - * - catch/unit/stream/hipStreamCreatePerformance.cc - * Test requirements - * ------------------------ - * - HIP_VERSION >= 5.6 - */ - -TEST_CASE("Unit_hipStreamCreate_Performance") { - HIP_CHECK(hipSetDevice(0)); // just to initialise HIP runtime. - // create stream - hipStream_t streamb1[4]; - // record time for batch1 stream creation - auto Startb1 = std::chrono::high_resolution_clock::now(); - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamCreate(&streamb1[k])); - } - auto Stopb1 = std::chrono::high_resolution_clock::now(); - double performb1 = std::chrono::duration(Stopb1 - Startb1).count(); // NOLINT - printf("Stream create performance for batch1 is %lf\n", performb1); - - // record time for batch2 stream creation before - // destroying already created streams - hipStream_t streamb2[4]; - auto Startb2 = std::chrono::high_resolution_clock::now(); - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamCreate(&streamb2[k])); - } - auto Stopb2 = std::chrono::high_resolution_clock::now(); - double performb2 = std::chrono::duration(Stopb2 - Startb2).count(); // NOLINT - printf("Stream create performance for batch2 is %lf\n", performb2); - // destroy batch 1 streams - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamDestroy(streamb1[k])); - } - // destroy batch 2 streams - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamDestroy(streamb2[k])); - } - // record time for batch3 stream creation after stream destroy - hipStream_t streamb3[4]; - auto Startb3 = std::chrono::high_resolution_clock::now(); - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamCreate(&streamb3[k])); - } - auto Stopb3 = std::chrono::high_resolution_clock::now(); - double performb3 = std::chrono::duration(Stopb3 - Startb3).count(); // NOLINT - printf("Stream create performance for batch3 is %lf\n", performb3); - - // destroy streams - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamDestroy(streamb3[k])); - } - REQUIRE(performb1 > performb2); - REQUIRE(performb2 > performb3); -} - -/** - * End doxygen group hipStreamCreate. - * @} - */ - -/** - * @addtogroup hipStreamCreateWithFlags hipStreamCreateWithFlags - * @{ - * @ingroup StreamTest - * `hipError_t hipStreamCreateWithFlags(hipStream_t* stream, unsigned int flags)` - - * Create a new asynchronous stream. - */ - -/** - * Test Description - * ------------------------ - * - Test case to verify hipStreamCreateWithFlags performance with - * hipStreamNonBlocking flagby recording below sets of time. - * create 4 set of streams and record that time taken as b1 - * create another 4 set of streams before destroying earlier streams - * and record time taken as b2 - * destroy streams and then create 4 streams again and record time taken as b3 - * and verify if the condition b1 > b2 > b3 is satisfied. - - * Test source - * ------------------------ - * - catch/unit/stream/hipStreamCreatePerformance.cc - * Test requirements - * ------------------------ - * - HIP_VERSION >= 5.6 - */ - -TEST_CASE("Unit_hipStreamCreate_WithFlagsPerformance_Nonblocking") { - HIP_CHECK(hipSetDevice(0)); // just to initialise HIP runtime. - // create stream - hipStream_t streamb1[4]; - // record time for batch1 stream creation - auto Startb1 = std::chrono::high_resolution_clock::now(); - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamCreateWithFlags(&streamb1[k], hipStreamNonBlocking)); - } - auto Stopb1 = std::chrono::high_resolution_clock::now(); - double performb1 = std::chrono::duration(Stopb1 - Startb1).count(); // NOLINT - printf("Stream create performance for batch1 is %lf\n", performb1); - - // record time for batch2 stream creation before - // destroying already created streams - hipStream_t streamb2[4]; - auto Startb2 = std::chrono::high_resolution_clock::now(); - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamCreateWithFlags(&streamb2[k], hipStreamNonBlocking)); - } - auto Stopb2 = std::chrono::high_resolution_clock::now(); - double performb2 = std::chrono::duration(Stopb2 - Startb2).count(); // NOLINT - printf("Stream create performance for batch2 is %lf\n", performb2); - // destroy batch 1 streams - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamDestroy(streamb1[k])); - } - // destroy batch 2 streams - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamDestroy(streamb2[k])); - } - // record time for batch3 stream creation after stream destroy - hipStream_t streamb3[4]; - auto Startb3 = std::chrono::high_resolution_clock::now(); - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamCreateWithFlags(&streamb3[k], hipStreamNonBlocking)); - } - auto Stopb3 = std::chrono::high_resolution_clock::now(); - double performb3 = std::chrono::duration(Stopb3 - Startb3).count(); // NOLINT - printf("Stream create performance for batch3 is %lf\n", performb3); - - // destroy streams - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamDestroy(streamb3[k])); - } - REQUIRE(performb1 > performb2); - REQUIRE(performb2 > performb3); -} - -/** - * Test Description - * ------------------------ - * - Test case to verify hipStreamCreateWithFlags performance - * with hipStreamDefault flagby recording below sets of time. - * create 4 set of streams and record that time taken as b1 - * create another 4 set of streams before destroying earlier streams - * and record time taken as b2 - * destroy streams and then create 4 streams again and record time taken as b3 - * and verify if the condition b1 > b2 > b3 is satisfied. - - * Test source - * ------------------------ - * - catch/unit/stream/hipStreamCreatePerformance.cc - * Test requirements - * ------------------------ - * - HIP_VERSION >= 5.6 - */ - -TEST_CASE("Unit_hipStreamCreate_WithFlagsPerformance_Default") { - HIP_CHECK(hipSetDevice(0)); // just to initialise HIP runtime. - // create stream - hipStream_t streamb1[4]; - // record time for batch1 stream creation - auto Startb1 = std::chrono::high_resolution_clock::now(); - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamCreateWithFlags(&streamb1[k], hipStreamDefault)); - } - auto Stopb1 = std::chrono::high_resolution_clock::now(); - double performb1 = std::chrono::duration(Stopb1 - Startb1).count(); // NOLINT - printf("Stream create performance for batch1 is %lf\n", performb1); - - // record time for batch2 stream creation before - // destroying already created streams - hipStream_t streamb2[4]; - auto Startb2 = std::chrono::high_resolution_clock::now(); - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamCreateWithFlags(&streamb2[k], hipStreamDefault)); - } - auto Stopb2 = std::chrono::high_resolution_clock::now(); - double performb2 = std::chrono::duration(Stopb2 - Startb2).count(); // NOLINT - printf("Stream create performance for batch2 is %lf\n", performb2); - // destroy batch 1 streams - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamDestroy(streamb1[k])); - } - // destroy batch 2 streams - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamDestroy(streamb2[k])); - } - // record time for batch3 stream creation after stream destroy - hipStream_t streamb3[4]; - auto Startb3 = std::chrono::high_resolution_clock::now(); - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamCreateWithFlags(&streamb3[k], hipStreamDefault)); - } - auto Stopb3 = std::chrono::high_resolution_clock::now(); - double performb3 = std::chrono::duration(Stopb3 - Startb3).count(); // NOLINT - printf("Stream create performance for batch3 is %lf\n", performb3); - - // destroy streams - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamDestroy(streamb3[k])); - } - REQUIRE(performb1 > performb2); - REQUIRE(performb2 > performb3); -} -/** - * End doxygen group hipStreamCreateWithFlags. - * @} - */ - -/** - * @addtogroup hipStreamCreateWithPriority hipStreamCreateWithPriority - * @{ - * @ingroup StreamTest - * `hipError_t hipStreamCreateWithPriority(hipStream_t* stream, unsigned int flags, int priority)` - - * Create a new asynchronous stream. - */ - -/** - * Test Description - * ------------------------ - * - Test case to verify hipStreamCreateWithPriority performance - * with hipStreamNonBlocking flag along with low priority - * by recording below sets of time. - * create 4 set of streams and record that time taken as b1 - * create another 4 set of streams before destroying earlier streams - * and record time taken as b2 - * destroy streams and then create 4 streams again and record time taken as b3 - * and verify if the condition b1 > b2 > b3 is satisfied. - - * Test source - * ------------------------ - * - catch/unit/stream/hipStreamCreatePerformance.cc - * Test requirements - * ------------------------ - * - HIP_VERSION >= 5.6 - */ - -TEST_CASE("Unit_hipStreamCreate_WithPriorityPerformance_Nonblocking_low") { - int priority_low, priority_high; - HIP_CHECK(hipSetDevice(0)); // just to initialise HIP runtime. - HIP_CHECK(hipDeviceGetStreamPriorityRange(&priority_low, &priority_high)); - // create stream - hipStream_t streamb1[4]; - // record time for batch1 stream creation - auto Startb1 = std::chrono::high_resolution_clock::now(); - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamCreateWithPriority(&streamb1[k], hipStreamNonBlocking, priority_low)); - } - auto Stopb1 = std::chrono::high_resolution_clock::now(); - double performb1 = std::chrono::duration(Stopb1 - Startb1).count(); // NOLINT - printf("Stream create performance for batch1 is %lf\n", performb1); - - // record time for batch2 stream creation before - // destroying already created streams - hipStream_t streamb2[4]; - auto Startb2 = std::chrono::high_resolution_clock::now(); - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamCreateWithPriority(&streamb2[k], hipStreamNonBlocking, priority_low)); - } - auto Stopb2 = std::chrono::high_resolution_clock::now(); - double performb2 = std::chrono::duration(Stopb2 - Startb2).count(); // NOLINT - printf("Stream create performance for batch2 is %lf\n", performb2); - // destroy batch 1 streams - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamDestroy(streamb1[k])); - } - // destroy batch 2 streams - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamDestroy(streamb2[k])); - } - // record time for batch3 stream creation after stream destroy - hipStream_t streamb3[4]; - auto Startb3 = std::chrono::high_resolution_clock::now(); - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamCreateWithPriority(&streamb3[k], hipStreamNonBlocking, priority_low)); - } - auto Stopb3 = std::chrono::high_resolution_clock::now(); - double performb3 = std::chrono::duration(Stopb3 - Startb3).count(); // NOLINT - printf("Stream create performance for batch3 is %lf\n", performb3); - - // destroy streams - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamDestroy(streamb3[k])); - } - REQUIRE(performb1 > performb2); - REQUIRE(performb2 > performb3); -} - -/** - * Test Description - * ------------------------ - * - Test case to verify hipStreamCreateWithPriority performance - * with hipStreamNonBlocking flag along with high priority - * by recording below sets of time. - * create 4 set of streams and record that time taken as b1 - * create another 4 set of streams before destroying earlier streams - * and record time taken as b2 - * destroy streams and then create 4 streams again and record time taken as b3 - * and verify if the condition b1 > b2 > b3 is satisfied. - - * Test source - * ------------------------ - * - catch/unit/stream/hipStreamCreatePerformance.cc - * Test requirements - * ------------------------ - * - HIP_VERSION >= 5.6 - */ - -TEST_CASE("Unit_hipStreamCreate_WithPriorityPerformance_Nonblocking_high") { - int priority_low, priority_high; - HIP_CHECK(hipSetDevice(0)); // just to initialise HIP runtime. - HIP_CHECK(hipDeviceGetStreamPriorityRange(&priority_low, &priority_high)); - // create stream - hipStream_t streamb1[4]; - // record time for batch1 stream creation - auto Startb1 = std::chrono::high_resolution_clock::now(); - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamCreateWithPriority(&streamb1[k], hipStreamNonBlocking, priority_high)); - } - auto Stopb1 = std::chrono::high_resolution_clock::now(); - double performb1 = std::chrono::duration(Stopb1 - Startb1).count(); // NOLINT - printf("Stream create performance for batch1 is %lf\n", performb1); - - // record time for batch2 stream creation before - // destroying already created streams - hipStream_t streamb2[4]; - auto Startb2 = std::chrono::high_resolution_clock::now(); - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamCreateWithPriority(&streamb2[k], hipStreamNonBlocking, priority_high)); - } - auto Stopb2 = std::chrono::high_resolution_clock::now(); - double performb2 = std::chrono::duration(Stopb2 - Startb2).count(); // NOLINT - printf("Stream create performance for batch2 is %lf\n", performb2); - // destroy batch 1 streams - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamDestroy(streamb1[k])); - } - // destroy batch 2 streams - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamDestroy(streamb2[k])); - } - // record time for batch3 stream creation after stream destroy - hipStream_t streamb3[4]; - auto Startb3 = std::chrono::high_resolution_clock::now(); - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamCreateWithPriority(&streamb3[k], hipStreamNonBlocking, priority_high)); - } - auto Stopb3 = std::chrono::high_resolution_clock::now(); - double performb3 = std::chrono::duration(Stopb3 - Startb3).count(); // NOLINT - printf("Stream create performance for batch3 is %lf\n", performb3); - - // destroy streams - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamDestroy(streamb3[k])); - } - REQUIRE(performb1 > performb2); - REQUIRE(performb2 > performb3); -} - -/** - * Test Description - * ------------------------ - * - Test case to verify hipStreamCreateWithPriority performance - * with hipStreamDefault flag along with low priority by recording below sets of time. - * create 4 set of streams and record that time taken as b1 - * create another 4 set of streams before destroying earlier streams - * and record time taken as b2 - * destroy streams and then create 4 streams again and record time taken as b3 - * and verify if the condition b1 > b2 > b3 is satisfied. - - * Test source - * ------------------------ - * - catch/unit/stream/hipStreamCreatePerformance.cc - * Test requirements - * ------------------------ - * - HIP_VERSION >= 5.6 - */ - -TEST_CASE("Unit_hipStreamCreate_WithPriorityPerformance_Default_low") { - int priority_low, priority_high; - HIP_CHECK(hipSetDevice(0)); // just to initialise HIP runtime. - HIP_CHECK(hipDeviceGetStreamPriorityRange(&priority_low, &priority_high)); - // create stream - hipStream_t streamb1[4]; - // record time for batch1 stream creation - auto Startb1 = std::chrono::high_resolution_clock::now(); - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamCreateWithPriority(&streamb1[k], hipStreamDefault, priority_low)); - } - auto Stopb1 = std::chrono::high_resolution_clock::now(); - double performb1 = std::chrono::duration(Stopb1 - Startb1).count(); // NOLINT - printf("Stream create performance for batch1 is %lf\n", performb1); - - // record time for batch2 stream creation before - // destroying already created streams - hipStream_t streamb2[4]; - auto Startb2 = std::chrono::high_resolution_clock::now(); - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamCreateWithPriority(&streamb2[k], hipStreamDefault, priority_low)); - } - auto Stopb2 = std::chrono::high_resolution_clock::now(); - double performb2 = std::chrono::duration(Stopb2 - Startb2).count(); // NOLINT - printf("Stream create performance for batch2 is %lf\n", performb2); - // destroy batch 1 streams - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamDestroy(streamb1[k])); - } - // destroy batch 2 streams - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamDestroy(streamb2[k])); - } - // record time for batch3 stream creation after stream destroy - hipStream_t streamb3[4]; - auto Startb3 = std::chrono::high_resolution_clock::now(); - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamCreateWithPriority(&streamb3[k], hipStreamDefault, priority_low)); - } - auto Stopb3 = std::chrono::high_resolution_clock::now(); - double performb3 = std::chrono::duration(Stopb3 - Startb3).count(); // NOLINT - printf("Stream create performance for batch3 is %lf\n", performb3); - - // destroy streams - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamDestroy(streamb3[k])); - } - REQUIRE(performb1 > performb2); - REQUIRE(performb2 > performb3); -} - -/** - * Test Description - * ------------------------ - * - Test case to verify hipStreamCreateWithPriority performance - * with hipStreamDefault flag along with high priority by recording below sets of time. - * create 4 set of streams and record that time taken as b1 - * create another 4 set of streams before destroying earlier streams - * and record time taken as b2 - * destroy streams and then create 4 streams again and record time taken as b3 - * and verify if the condition b1 > b2 > b3 is satisfied. - - * Test source - * ------------------------ - * - catch/unit/stream/hipStreamCreatePerformance.cc - * Test requirements - * ------------------------ - * - HIP_VERSION >= 5.6 - */ - -TEST_CASE("Unit_hipStreamCreate_WithPriorityPerformance_Default_high") { - int priority_low = 1, priority_high = 2; - HIP_CHECK(hipSetDevice(0)); // just to initialise HIP runtime. - HIP_CHECK(hipDeviceGetStreamPriorityRange(&priority_low, &priority_high)); - // create stream - hipStream_t streamb1[4]; - // record time for batch1 stream creation - auto Startb1 = std::chrono::high_resolution_clock::now(); - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamCreateWithPriority(&streamb1[k], hipStreamDefault, priority_high)); - } - auto Stopb1 = std::chrono::high_resolution_clock::now(); - double performb1 = std::chrono::duration(Stopb1 - Startb1).count(); // NOLINT - printf("Stream create performance for batch1 is %lf\n", performb1); - - // record time for batch2 stream creation before - // destroying already created streams - hipStream_t streamb2[4]; - auto Startb2 = std::chrono::high_resolution_clock::now(); - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamCreateWithPriority(&streamb2[k], hipStreamDefault, priority_high)); - } - auto Stopb2 = std::chrono::high_resolution_clock::now(); - double performb2 = std::chrono::duration(Stopb2 - Startb2).count(); // NOLINT - printf("Stream create performance for batch2 is %lf\n", performb2); - // destroy batch 1 streams - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamDestroy(streamb1[k])); - } - // destroy batch 2 streams - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamDestroy(streamb2[k])); - } - // record time for batch3 stream creation after stream destroy - hipStream_t streamb3[4]; - auto Startb3 = std::chrono::high_resolution_clock::now(); - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamCreateWithPriority(&streamb3[k], hipStreamDefault, priority_high)); - } - auto Stopb3 = std::chrono::high_resolution_clock::now(); - double performb3 = std::chrono::duration(Stopb3 - Startb3).count(); // NOLINT - printf("Stream create performance for batch3 is %lf\n", performb3); - - // destroy streams - for (int k = 0; k < 4; k++) { - HIP_CHECK(hipStreamDestroy(streamb3[k])); - } - REQUIRE(performb1 > performb2); - REQUIRE(performb2 > performb3); -} -/** - * End doxygen group hipStreamCreateWithPriority. - * @} - */