From 1ebac1664eaa6099ad7d3c251f134efea299929e Mon Sep 17 00:00:00 2001 From: =?UTF-8?q?Mirza=20Halil=C4=8Devi=C4=87?= <109971222+mirza-halilcevic@users.noreply.github.com> Date: Thu, 1 Dec 2022 05:37:20 +0100 Subject: [PATCH] EXSWHTEC-3 - Implement tests for hipDeviceReset (#2873) * Implement tests for hipDeviceReset - Basic test for behaviour validation - Multithreaded test similar to basic test. * Update license boilerplate. Signed-off-by: Dino Music * EXSWHTEC-3 - Add stream and memory checks to hipDeviceReset tests. * EXSWHTEC-3 - Corrected expected error code for hipStreamDestroy * EXSWHTEC-3 - Corrected check ordering in Unit_hipDeviceReset_Positive_Threaded test. * EXSWHTEC-3 - Exclude hipDeviceSetGetCacheConfig checks for AMD * EXSWHTEC-3 - Corrected merge error in device/CMakeLists.txt * EXSWHTEC-3 - Remove NVIDIA guards from tests and relax the code. * EXSWHTEC-3 - Update tests to take into account unsupported operations on AMD * EXSWHTEC-3 - Fix bug in Unit_hipDeviceReset_Positive_Threaded test. * EXSWHTEC-3 - Disable hipDeviceReset tests for AMD. [ROCm/hip-tests commit: 65ba7da60733b3f7c3845bf51daf896c8f24f337] --- .../config/config_amd_linux_common.json | 4 +- .../config/config_amd_windows_common.json | 2 + .../catch/unit/device/CMakeLists.txt | 1 + .../catch/unit/device/hipDeviceReset.cc | 131 ++++++++++++++++++ 4 files changed, 137 insertions(+), 1 deletion(-) create mode 100644 projects/hip-tests/catch/unit/device/hipDeviceReset.cc diff --git a/projects/hip-tests/catch/hipTestMain/config/config_amd_linux_common.json b/projects/hip-tests/catch/hipTestMain/config/config_amd_linux_common.json index ccb77bc2a4..742c45dcf1 100644 --- a/projects/hip-tests/catch/hipTestMain/config/config_amd_linux_common.json +++ b/projects/hip-tests/catch/hipTestMain/config/config_amd_linux_common.json @@ -12,6 +12,8 @@ "Unit_hipIpcOpenMemHandle_Negative_Open_In_Creating_Process", "Unit_hipDeviceGetPCIBusId_Negative_PartialFill", "Unit_hipInit_Negative", - "Unit_hipMemset_Negative_OutOfBoundsPtr" + "Unit_hipMemset_Negative_OutOfBoundsPtr", + "Unit_hipDeviceReset_Positive_Basic", + "Unit_hipDeviceReset_Positive_Threaded" ] } diff --git a/projects/hip-tests/catch/hipTestMain/config/config_amd_windows_common.json b/projects/hip-tests/catch/hipTestMain/config/config_amd_windows_common.json index 33da571b15..4724985967 100644 --- a/projects/hip-tests/catch/hipTestMain/config/config_amd_windows_common.json +++ b/projects/hip-tests/catch/hipTestMain/config/config_amd_windows_common.json @@ -80,6 +80,8 @@ "Unit_hipDeviceGetPCIBusId_Negative_PartialFill", "Unit_hipDeviceGetSharedMemConfig_Positive_Basic", "Unit_hipDeviceGetSharedMemConfig_Positive_Threaded", + "Unit_hipDeviceReset_Positive_Basic", + "Unit_hipDeviceReset_Positive_Threaded", "Unit_hipInit_Negative", "Unit_hipGraphMemcpyNodeSetParams_Functional", "Unit_hipGraphNodeGetDependentNodes_Functional", diff --git a/projects/hip-tests/catch/unit/device/CMakeLists.txt b/projects/hip-tests/catch/unit/device/CMakeLists.txt index 9f11be6877..1e26944d0f 100644 --- a/projects/hip-tests/catch/unit/device/CMakeLists.txt +++ b/projects/hip-tests/catch/unit/device/CMakeLists.txt @@ -23,6 +23,7 @@ set(TEST_SRC hipExtGetLinkTypeAndHopCount.cc hipDeviceSetLimit.cc hipDeviceSetGetSharedMemConfig.cc + hipDeviceReset.cc hipDeviceSetGetMemPool.cc hipInit.cc hipDriverGetVersion.cc diff --git a/projects/hip-tests/catch/unit/device/hipDeviceReset.cc b/projects/hip-tests/catch/unit/device/hipDeviceReset.cc new file mode 100644 index 0000000000..109121612a --- /dev/null +++ b/projects/hip-tests/catch/unit/device/hipDeviceReset.cc @@ -0,0 +1,131 @@ +/* +Copyright (c) 2022 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 + +TEST_CASE("Unit_hipDeviceReset_Positive_Basic") { + const auto device = GENERATE(range(0, HipTest::getDeviceCount())); + HIP_CHECK(hipSetDevice(device)); + INFO("Current device is: " << device); + + unsigned int flags_before = 0u; + HIP_CHECK(hipGetDeviceFlags(&flags_before)); + hipSharedMemConfig mem_config_before; + HIP_CHECK(hipDeviceGetSharedMemConfig(&mem_config_before)); + + void* ptr = nullptr; + HIP_CHECK(hipMalloc(&ptr, 500)); + hipStream_t stream = nullptr; + HIP_CHECK(hipStreamCreate(&stream)); + + const auto cache_config_ret = hipDeviceSetCacheConfig(hipFuncCachePreferL1); + REQUIRE((cache_config_ret == hipSuccess || cache_config_ret == hipErrorNotSupported)); + + const auto shared_mem_config_ret = hipDeviceSetSharedMemConfig( + mem_config_before == hipSharedMemBankSizeFourByte ? hipSharedMemBankSizeEightByte + : hipSharedMemBankSizeFourByte); + REQUIRE((shared_mem_config_ret == hipSuccess || shared_mem_config_ret == hipErrorNotSupported)); + + HIP_CHECK(hipSetDeviceFlags(flags_before ^ (1u << 2))); + + HIP_CHECK(hipDeviceReset()); + + unsigned int flags_after = 0u; + CHECK(hipGetDeviceFlags(&flags_after) == hipSuccess); + CHECK(flags_after == flags_before); + + CHECK(hipFree(ptr) == hipErrorInvalidValue); + +// Inconsistent behavior in CUDA, sometimes segfaults, sometimes works +// Return value mismatch on AMD - EXSWHTEC-124 +#if HT_NVIDIA + CHECK(hipStreamDestroy(stream) == hipErrorInvalidHandle); +#endif + + if (cache_config_ret == hipSuccess) { + hipFuncCache_t cache_config; + CHECK(hipDeviceGetCacheConfig(&cache_config) == hipSuccess); + CHECK(cache_config == hipFuncCachePreferNone); + } + + if (shared_mem_config_ret == hipSuccess) { + hipSharedMemConfig mem_config_after; + CHECK(hipDeviceGetSharedMemConfig(&mem_config_after) == hipSuccess); + CHECK(mem_config_after == mem_config_before); + } +} + +TEST_CASE("Unit_hipDeviceReset_Positive_Threaded") { + HIP_CHECK(hipSetDevice(0)); + INFO("Current device is: " << 0); + + unsigned int flags_before = 0u; + HIP_CHECK(hipGetDeviceFlags(&flags_before)); + hipSharedMemConfig mem_config_before; + HIP_CHECK(hipDeviceGetSharedMemConfig(&mem_config_before)); + + void* ptr = nullptr; + HIP_CHECK(hipMalloc(&ptr, 500)); + hipStream_t stream = nullptr; + HIP_CHECK(hipStreamCreate(&stream)); + + const auto cache_config_ret = hipDeviceSetCacheConfig(hipFuncCachePreferL1); + REQUIRE((cache_config_ret == hipSuccess || cache_config_ret == hipErrorNotSupported)); + + const auto shared_mem_config_ret = hipDeviceSetSharedMemConfig( + mem_config_before == hipSharedMemBankSizeFourByte ? hipSharedMemBankSizeEightByte + : hipSharedMemBankSizeFourByte); + REQUIRE((shared_mem_config_ret == hipSuccess || shared_mem_config_ret == hipErrorNotSupported)); + + + HIP_CHECK(hipSetDeviceFlags(flags_before ^ (1u << 2))); + + std::thread([] { + HIP_CHECK_THREAD(hipSetDevice(0)); + HIP_CHECK_THREAD(hipDeviceReset()); + }).join(); + HIP_CHECK_THREAD_FINALIZE(); + + unsigned int flags_after = 0u; + CHECK(hipGetDeviceFlags(&flags_after) == hipSuccess); + CHECK(flags_after == flags_before); + + CHECK(hipFree(ptr) == hipErrorInvalidValue); + +// Inconsistent behavior in CUDA, sometimes segfaults, sometimes works +// Return value mismatch on AMD - EXSWHTEC-124 +#if HT_NVIDIA + CHECK(hipStreamDestroy(stream) == hipErrorInvalidHandle); +#endif + + if (cache_config_ret == hipSuccess) { + hipFuncCache_t cache_config; + CHECK(hipDeviceGetCacheConfig(&cache_config) == hipSuccess); + CHECK(cache_config == hipFuncCachePreferNone); + } + + if (shared_mem_config_ret == hipSuccess) { + hipSharedMemConfig mem_config_after; + CHECK(hipDeviceGetSharedMemConfig(&mem_config_after) == hipSuccess); + CHECK(mem_config_after == mem_config_before); + } +} \ No newline at end of file