From a47e8ac84fe5dd3174ba1a9ebce8def3521c3491 Mon Sep 17 00:00:00 2001 From: ROCm CI Service Account <66695075+rocm-ci@users.noreply.github.com> Date: Sat, 5 Nov 2022 04:18:13 +0530 Subject: [PATCH 1/6] SWDEV-362930 - Fixes packging issue Windows using makefile (#3054) Issue NMAKE : fatal error U1065: invalid option 'w' Solution - Windows to use Ninja but cmake needs to skip compiler test Change-Id: I1f666254ce5188dff0897815677ce4d9faf55e4c --- tests/catch/packaging/hip-tests.txt | 7 +++++++ 1 file changed, 7 insertions(+) diff --git a/tests/catch/packaging/hip-tests.txt b/tests/catch/packaging/hip-tests.txt index 553fdb853b..3db1512dfe 100644 --- a/tests/catch/packaging/hip-tests.txt +++ b/tests/catch/packaging/hip-tests.txt @@ -19,6 +19,13 @@ # OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN # THE SOFTWARE. cmake_minimum_required(VERSION 3.16.8) +# vc 19.31.31107.0 gives issue while packaging using makefile +# To avoid error NMAKE : fatal error U1065: invalid option 'w' +# Windows to use Ninja generator like other projects +# to skip the simple compiler test +set(CMAKE_C_COMPILER_WORKS 1) +set(CMAKE_CXX_COMPILER_WORKS 1) + project(tests) MACRO(SUBDIRLIST result curdir) From a9e9d9f99648a365ef81a44adeeb5b53f4db6796 Mon Sep 17 00:00:00 2001 From: ROCm CI Service Account <66695075+rocm-ci@users.noreply.github.com> Date: Sat, 5 Nov 2022 05:44:41 +0530 Subject: [PATCH 2/6] SWDEV-354905 - prioritize -DROCM_PATH over env{ROCM_PATH} in hip tests (#3055) Change-Id: I943bc705641ceccda95b409d13b4e757c4e5ab23 --- tests/catch/CMakeLists.txt | 7 +++++-- 1 file changed, 5 insertions(+), 2 deletions(-) diff --git a/tests/catch/CMakeLists.txt b/tests/catch/CMakeLists.txt index e2924a561d..3f6c992059 100644 --- a/tests/catch/CMakeLists.txt +++ b/tests/catch/CMakeLists.txt @@ -57,13 +57,16 @@ else() # Set both compilers else windows cmake complains of mismatch cmake_path(SET CMAKE_CXX_COMPILER "${HIP_PATH}/bin/hipcc.bat") cmake_path(SET CMAKE_C_COMPILER "${HIP_PATH}/bin/hipcc.bat") - set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} --rocm-path=${ROCM_PATH}") set(HIPCONFIG_EXECUTABLE "${HIP_PATH}/bin/hipconfig.bat") execute_process(COMMAND ${HIPCONFIG_EXECUTABLE} --version OUTPUT_VARIABLE HIP_VERSION OUTPUT_STRIP_TRAILING_WHITESPACE) endif() -# enforce c++17 for all tests +if(HIP_PLATFORM STREQUAL "amd") + # prioritize -DROCM_PATH over env{ROCM_PATH} for amd platform only + set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} --rocm-path=${ROCM_PATH}") +endif() +# enforce c++17 set(CMAKE_CXX_FLAGS "${CMAKE_CXX_FLAGS} --std=c++17") string(REPLACE "." ";" VERSION_LIST ${HIP_VERSION}) From c504292c5411071439cfcd961ccc47b72c389093 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: Mon, 7 Nov 2022 09:31:10 +0100 Subject: [PATCH 3/6] EXSWHTEC-12 - Implement additional tests for hipIpcGetMemHandle (#2879) - Verify that unique handles are returned for consecutive calls - Verify that unique handles are returned for separate allocations at same address - Verify that freed memory causes an error - Verify that out of bounds pointer will cause an error. - Remove linux guard from test file and exclude test from windows build with cmake. - Fix expected error in hipIpcGetMemHandle negative tests. --- tests/catch/unit/device/CMakeLists.txt | 4 +- tests/catch/unit/device/hipIpcGetMemHandle.cc | 73 +++++++++++++++++++ 2 files changed, 76 insertions(+), 1 deletion(-) create mode 100644 tests/catch/unit/device/hipIpcGetMemHandle.cc diff --git a/tests/catch/unit/device/CMakeLists.txt b/tests/catch/unit/device/CMakeLists.txt index 4b98e35fb0..1ac6b88fbb 100644 --- a/tests/catch/unit/device/CMakeLists.txt +++ b/tests/catch/unit/device/CMakeLists.txt @@ -27,7 +27,9 @@ set(TEST_SRC if(UNIX) set(TEST_SRC ${TEST_SRC} - hipIpcCloseMemHandle.cc) + hipIpcGetMemHandle.cc + hipIpcCloseMemHandle.cc + ) endif() set_source_files_properties(hipGetDeviceCount.cc PROPERTIES COMPILE_FLAGS -std=c++17) diff --git a/tests/catch/unit/device/hipIpcGetMemHandle.cc b/tests/catch/unit/device/hipIpcGetMemHandle.cc new file mode 100644 index 0000000000..76d2c92ffc --- /dev/null +++ b/tests/catch/unit/device/hipIpcGetMemHandle.cc @@ -0,0 +1,73 @@ +/* +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 +#include + +TEST_CASE("Unit_hipIpcGetMemHandle_Positive_Unique_Handles_Separate_Allocations") { + void *ptr1, *ptr2; + hipIpcMemHandle_t handle1, handle2; + HIP_CHECK(hipMalloc(&ptr1, 1024)); + HIP_CHECK(hipMalloc(&ptr2, 1024)); + HIP_CHECK(hipIpcGetMemHandle(&handle1, ptr1)); + HIP_CHECK(hipIpcGetMemHandle(&handle2, ptr2)); + + CHECK(memcmp(&handle1, &handle2, sizeof(handle1)) != 0); + + HIP_CHECK(hipFree(ptr1)); + HIP_CHECK(hipFree(ptr2)); +} + +TEST_CASE("Unit_hipIpcGetMemHandle_Positive_Unique_Handles_Reused_Memory") { + void *ptr1 = nullptr, *ptr2 = nullptr; + hipIpcMemHandle_t handle1, handle2; + HIP_CHECK(hipMalloc(&ptr1, 1024)); + HIP_CHECK(hipIpcGetMemHandle(&handle1, ptr1)); + HIP_CHECK(hipFree(ptr1)); + + HIP_CHECK(hipMalloc(&ptr2, 1024)); + HIP_CHECK(hipIpcGetMemHandle(&handle2, ptr2)); + + if (ptr1 == ptr2) CHECK(memcmp(&handle1, &handle2, sizeof(handle1)) != 0); + + HIP_CHECK(hipFree(ptr2)); +} + +TEST_CASE("Unit_hipIpcGetMemHandle_Negative_Handle_For_Freed_Memory") { + void* ptr; + hipIpcMemHandle_t handle; + HIP_CHECK(hipMalloc(&ptr, 1024)); + HIP_CHECK(hipFree(ptr)); + HIP_CHECK_ERROR(hipIpcGetMemHandle(&handle, ptr), hipErrorInvalidValue); +} + +TEST_CASE("Unit_hipIpcGetMemHandle_Negative_Out_Of_Bound_Pointer") { + int* ptr; + constexpr size_t n = 1024; + hipIpcMemHandle_t handle; + HIP_CHECK(hipMalloc(reinterpret_cast(&ptr), n * sizeof(*ptr))); + HIP_CHECK_ERROR(hipIpcGetMemHandle(&handle, reinterpret_cast(ptr + n)), + hipErrorInvalidValue); + HIP_CHECK(hipFree(reinterpret_cast(ptr))); +} \ No newline at end of file From cd4a634935b05024991660a152a079aff8e0f7ff 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: Mon, 7 Nov 2022 09:31:36 +0100 Subject: [PATCH 4/6] EXSWHTEC-27 - Implement tests for hipDeviceSetSharedMemConfig and hipDeviceGetSharedMemConfig (#2920) - Negative parameter tests - Default return value test - Multithreaded test - Adjust hipDeviceSetGetSharedMemConfig tests according to new return value. --- .../config/config_amd_linux_common.json | 2 + .../config/config_amd_windows_common.json | 4 +- tests/catch/unit/device/CMakeLists.txt | 1 + .../device/hipDeviceSetGetSharedMemConfig.cc | 121 ++++++++++++++++++ 4 files changed, 127 insertions(+), 1 deletion(-) create mode 100644 tests/catch/unit/device/hipDeviceSetGetSharedMemConfig.cc diff --git a/tests/catch/hipTestMain/config/config_amd_linux_common.json b/tests/catch/hipTestMain/config/config_amd_linux_common.json index 2a857c27d0..dd0f6fd6c4 100644 --- a/tests/catch/hipTestMain/config/config_amd_linux_common.json +++ b/tests/catch/hipTestMain/config/config_amd_linux_common.json @@ -3,6 +3,8 @@ [ "Unit_hipStreamPerThread_DeviceReset_1", "Unit_hipMallocManaged_OverSubscription", + "Unit_hipDeviceGetSharedMemConfig_Positive_Basic", + "Unit_hipDeviceGetSharedMemConfig_Positive_Threaded", "Unit_hipDeviceGetCacheConfig_Positive_Basic", "Unit_hipDeviceGetCacheConfig_Positive_Threaded", "Unit_hipGetDeviceFlags_Positive_Context", diff --git a/tests/catch/hipTestMain/config/config_amd_windows_common.json b/tests/catch/hipTestMain/config/config_amd_windows_common.json index 3c8570f63d..9662edd8ed 100644 --- a/tests/catch/hipTestMain/config/config_amd_windows_common.json +++ b/tests/catch/hipTestMain/config/config_amd_windows_common.json @@ -100,6 +100,8 @@ "Unit_hipStreamValue_Wait64_Blocking_NoMask_Nor", "Unit_hipGetDeviceFlags_Positive_Context", "Unit_hipIpcCloseMemHandle_Negative_Close_In_Originating_Process", - "Unit_hipDeviceGetPCIBusId_Negative_PartialFill" + "Unit_hipDeviceGetPCIBusId_Negative_PartialFill", + "Unit_hipDeviceGetSharedMemConfig_Positive_Basic", + "Unit_hipDeviceGetSharedMemConfig_Positive_Threaded" ] } diff --git a/tests/catch/unit/device/CMakeLists.txt b/tests/catch/unit/device/CMakeLists.txt index 1ac6b88fbb..f027abbfa7 100644 --- a/tests/catch/unit/device/CMakeLists.txt +++ b/tests/catch/unit/device/CMakeLists.txt @@ -22,6 +22,7 @@ set(TEST_SRC hipDeviceEnableDisablePeerAccess.cc hipExtGetLinkTypeAndHopCount.cc hipDeviceSetLimit.cc + hipDeviceSetGetSharedMemConfig.cc hipDeviceSetGetMemPool.cc ) diff --git a/tests/catch/unit/device/hipDeviceSetGetSharedMemConfig.cc b/tests/catch/unit/device/hipDeviceSetGetSharedMemConfig.cc new file mode 100644 index 0000000000..045f40092a --- /dev/null +++ b/tests/catch/unit/device/hipDeviceSetGetSharedMemConfig.cc @@ -0,0 +1,121 @@ +/* +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 +#include + +namespace { +constexpr std::array kMemConfigs{ + hipSharedMemBankSizeDefault, hipSharedMemBankSizeFourByte, hipSharedMemBankSizeEightByte}; +} // anonymous namespace + +TEST_CASE("Unit_hipDeviceSetSharedMemConfig_Positive_Basic") { + const auto device = GENERATE(range(0, HipTest::getDeviceCount())); + const auto mem_config = GENERATE(from_range(std::begin(kMemConfigs), std::end(kMemConfigs))); + HIP_CHECK(hipSetDevice(device)); + INFO("Current device is " << device); + +#if HT_AMD + HIP_CHECK_ERROR(hipDeviceSetSharedMemConfig(mem_config), hipErrorNotSupported); +#elif HT_NVIDIA + HIP_CHECK(hipDeviceSetSharedMemConfig(mem_config)); +#endif +} + +TEST_CASE("Unit_hipDeviceSetSharedMemConfig_Negative_Parameters") { +#if HT_AMD + HIP_CHECK_ERROR(hipDeviceSetSharedMemConfig(static_cast(-1)), + hipErrorNotSupported); +#elif HT_NVIDIA + HIP_CHECK_ERROR(hipDeviceSetSharedMemConfig(static_cast(-1)), + hipErrorInvalidValue); +#endif +} + +TEST_CASE("Unit_hipDeviceGetSharedMemConfig_Positive_Default") { + const auto device = GENERATE(range(0, HipTest::getDeviceCount())); + HIP_CHECK(hipSetDevice(device)); + INFO("Current device is " << device); + + hipSharedMemConfig mem_config; + HIP_CHECK(hipDeviceGetSharedMemConfig(&mem_config)); + REQUIRE(mem_config == hipSharedMemBankSizeFourByte); +} + +TEST_CASE("Unit_hipDeviceGetSharedMemConfig_Positive_Basic") { + const auto device = GENERATE(range(0, HipTest::getDeviceCount())); + const auto mem_config = GENERATE(from_range(std::begin(kMemConfigs), std::end(kMemConfigs))); + HIP_CHECK(hipSetDevice(device)); + INFO("Current device is " << device); + + HIP_CHECK(hipDeviceSetSharedMemConfig(mem_config)); + + hipSharedMemConfig returned_mem_config; + HIP_CHECK(hipDeviceGetSharedMemConfig(&returned_mem_config)); + + int major = -1, minor = -1; + HIP_CHECK(hipDeviceComputeCapability(&major, &minor, device)); + REQUIRE(major > 0); + if (major == 3 /*Kepler*/) { + REQUIRE(returned_mem_config == mem_config); + } else { + REQUIRE(returned_mem_config == hipSharedMemBankSizeFourByte); + } +} + +TEST_CASE("Unit_hipDeviceGetSharedMemConfig_Positive_Threaded") { + class HipDeviceGetSharedMemConfigTest + : public ThreadedZigZagTest { + public: + HipDeviceGetSharedMemConfigTest(const hipSharedMemConfig mem_config) + : mem_config_{mem_config} {} + + void TestPart2() { HIP_CHECK_THREAD(hipDeviceSetSharedMemConfig(mem_config_)); } + + void TestPart3() { + hipSharedMemConfig returned_mem_config; + HIP_CHECK(hipDeviceGetSharedMemConfig(&returned_mem_config)); + + int major = -1, minor = -1; + HIP_CHECK(hipDeviceComputeCapability(&major, &minor, 0)); + REQUIRE(major > 0); + if (major == 3 /*Kepler*/) { + REQUIRE(returned_mem_config == mem_config_); + } else { + REQUIRE(returned_mem_config == hipSharedMemBankSizeFourByte); + } + } + + private: + const hipSharedMemConfig mem_config_; + }; + + const auto mem_config = GENERATE(from_range(std::begin(kMemConfigs), std::end(kMemConfigs))); + + HipDeviceGetSharedMemConfigTest test(mem_config); + test.run(); +} + +TEST_CASE("Unit_hipDeviceGetSharedMemConfig_Negative_Parameters") { + HIP_CHECK_ERROR(hipDeviceGetSharedMemConfig(nullptr), hipErrorInvalidValue); +} \ No newline at end of file From 4fbc08e6c4a8aab7bec490ee455e57794e545dda Mon Sep 17 00:00:00 2001 From: ROCm CI Service Account <66695075+rocm-ci@users.noreply.github.com> Date: Mon, 7 Nov 2022 15:36:09 +0530 Subject: [PATCH 5/6] SWDEV-327563 - Windows Enable fixed tests v5 (#3059) Change-Id: Ie144e481f38b3e5d462f64d76ab4fd32043f2b46 --- .../config/config_amd_windows_common.json | 24 ------------------- tests/catch/unit/event/hipEventDestroy.cc | 5 ---- tests/catch/unit/stream/hipStreamDestroy.cc | 7 +----- 3 files changed, 1 insertion(+), 35 deletions(-) diff --git a/tests/catch/hipTestMain/config/config_amd_windows_common.json b/tests/catch/hipTestMain/config/config_amd_windows_common.json index 9662edd8ed..b0c106dbf3 100644 --- a/tests/catch/hipTestMain/config/config_amd_windows_common.json +++ b/tests/catch/hipTestMain/config/config_amd_windows_common.json @@ -10,7 +10,6 @@ "Unit_hipMemPoolApi_BasicTrim", "Unit_hipMemPoolApi_BasicReuse", "Unit_hipMemPoolApi_Opportunistic", - "Unit_hipGraphMemcpyNodeSetParams_Functional", "Unit_hipMalloc3D_ValidatePitch", "Unit_hipMemAllocPitch_ValidatePitch", "Unit_hipGraphExecMemcpyNodeSetParamsToSymbol_Functional", @@ -37,8 +36,6 @@ "Unit_hipMallocManaged_DeviceContextChange - int", "Unit_hipMallocManaged_DeviceContextChange - float", "Unit_hipMallocManaged_DeviceContextChange - double", - "Unit_hipGraphNodeGetDependentNodes_Functional", - "Unit_hipGraphNodeGetDependencies_Functional", "Unit_hipStreamCreateWithPriority_ValidateWithEvents", "Unit_hipStreamPerThread_StrmWaitEvt", "Unit_hipMemGetInfo_DifferentMallocSmall", @@ -51,32 +48,11 @@ "Unit_hipMemGetInfo_Malloc3DArray - int4", "Unit_hipMemGetInfo_ParaSmall", "Unit_hipMemGetInfo_ParaMultiSmall", - "Unit_hipFreeMultiTDev - char", - "Unit_hipFreeMultiTDev - int", - "Unit_hipFreeMultiTDev - float2", - "Unit_hipFreeMultiTDev - float4", - "Unit_hipFreeMultiTHost - char", - "Unit_hipFreeMultiTHost - int", - "Unit_hipFreeMultiTHost - float2", - "Unit_hipFreeMultiTHost - float4", - "Unit_hipFreeMultiTArray - char", - "Unit_hipFreeMultiTArray - int", - "Unit_hipFreeMultiTArray - float2", - "Unit_hipFreeMultiTArray - float4", - "Unit_hipStreamSynchronize_FinishWork", - "Unit_hipStreamSynchronize_NullStreamAndStreamPerThread", "Unit_hipMultiThreadDevice_NearZero", "Unit_hipStreamPerThread_DeviceReset_1", "Unit_hipStreamCreateWithPriority_ValidateWithEvents", "Unit_hipStreamPerThread_StrmWaitEvt", - "Unit_hipGraphHostNodeGetParams_ClonedGraphwithHostNode", - "Unit_hipGraphHostNodeGetParams_BasicFunc", - "Unit_hipGraphHostNodeGetParams_SetParams", - "Unit_hipGraphExecChildGraphNodeSetParams_BasicFunc", - "Unit_hipGraphExecChildGraphNodeSetParams_ChildTopology", - "Unit_hipGraphMemcpyNodeSetParams1D_Functional", "Unit_hipGraphMemcpyNodeSetParamsToSymbol_Functional", - "Unit_hipStreamQuery_WithPendingWork", "Unit_hipStreamWaitEvent_DifferentStreams", "Unit_hipStreamQuery_WithFinishedWork", "Unit_hipDeviceGetCacheConfig_Positive_Basic", diff --git a/tests/catch/unit/event/hipEventDestroy.cc b/tests/catch/unit/event/hipEventDestroy.cc index 23cb17bed5..8965d5227a 100644 --- a/tests/catch/unit/event/hipEventDestroy.cc +++ b/tests/catch/unit/event/hipEventDestroy.cc @@ -95,11 +95,6 @@ TEST_CASE("Unit_hipEventDestroy_WithWaitingStream") { } TEST_CASE("Unit_hipEventDestroy_Negative") { -#if HT_AMD - HipTest::HIP_SKIP_TEST("EXSWCPHIPT-103"); - return; -#endif - SECTION("Invalid Event") { hipEvent_t event{nullptr}; HIP_CHECK_ERROR(hipEventDestroy(event), hipErrorInvalidResourceHandle); diff --git a/tests/catch/unit/stream/hipStreamDestroy.cc b/tests/catch/unit/stream/hipStreamDestroy.cc index 54edddc7c6..7c7658e64b 100644 --- a/tests/catch/unit/stream/hipStreamDestroy.cc +++ b/tests/catch/unit/stream/hipStreamDestroy.cc @@ -72,12 +72,7 @@ TEST_CASE("Unit_hipStreamDestroy_WithFinishedWork") { // of work #if HT_AMD /* Disabled because frequency based wait is timing out on nvidia platforms */ TEST_CASE("Unit_hipStreamDestroy_WithPendingWork") { -#if HT_AMD - HipTest::HIP_SKIP_TEST( - "EXSWCPHIPT-44 - expected hipStreamDestroy to return immediately then release the resources " - "when the queued jobs are finished"); - return; -#endif + hipStream_t stream{}; HIP_CHECK(hipStreamCreate(&stream)); constexpr int numDataPoints = 10; From 809149ecc8d751acd3c1595b590090cd86ada8df 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: Mon, 7 Nov 2022 12:20:44 +0100 Subject: [PATCH 6/6] EXSWHTEC-10 - Implement additional tests for hipIpcOpenMemHandle (#2880) - Verify that opening the same handle in different contexts on same device causes an error - Verify that opening the handle in the process that created it causes an error - Remove linux guard from test file and exclude test from windows build with cmake. - Disable Unit_hipIpcOpenMemHandle_Negative_Open_In_Creating_Process for AMD. --- .../config/config_amd_linux_common.json | 1 + .../config/config_amd_windows_common.json | 1 + tests/catch/unit/device/CMakeLists.txt | 1 + .../catch/unit/device/hipIpcOpenMemHandle.cc | 88 +++++++++++++++++++ 4 files changed, 91 insertions(+) create mode 100644 tests/catch/unit/device/hipIpcOpenMemHandle.cc diff --git a/tests/catch/hipTestMain/config/config_amd_linux_common.json b/tests/catch/hipTestMain/config/config_amd_linux_common.json index dd0f6fd6c4..b3a7503222 100644 --- a/tests/catch/hipTestMain/config/config_amd_linux_common.json +++ b/tests/catch/hipTestMain/config/config_amd_linux_common.json @@ -9,6 +9,7 @@ "Unit_hipDeviceGetCacheConfig_Positive_Threaded", "Unit_hipGetDeviceFlags_Positive_Context", "Unit_hipIpcCloseMemHandle_Negative_Close_In_Originating_Process", + "Unit_hipIpcOpenMemHandle_Negative_Open_In_Creating_Process", "Unit_hipDeviceGetPCIBusId_Negative_PartialFill" ] } diff --git a/tests/catch/hipTestMain/config/config_amd_windows_common.json b/tests/catch/hipTestMain/config/config_amd_windows_common.json index b0c106dbf3..17ae9eb145 100644 --- a/tests/catch/hipTestMain/config/config_amd_windows_common.json +++ b/tests/catch/hipTestMain/config/config_amd_windows_common.json @@ -76,6 +76,7 @@ "Unit_hipStreamValue_Wait64_Blocking_NoMask_Nor", "Unit_hipGetDeviceFlags_Positive_Context", "Unit_hipIpcCloseMemHandle_Negative_Close_In_Originating_Process", + "Unit_hipIpcOpenMemHandle_Negative_Open_In_Creating_Process", "Unit_hipDeviceGetPCIBusId_Negative_PartialFill", "Unit_hipDeviceGetSharedMemConfig_Positive_Basic", "Unit_hipDeviceGetSharedMemConfig_Positive_Threaded" diff --git a/tests/catch/unit/device/CMakeLists.txt b/tests/catch/unit/device/CMakeLists.txt index f027abbfa7..e8a738c56e 100644 --- a/tests/catch/unit/device/CMakeLists.txt +++ b/tests/catch/unit/device/CMakeLists.txt @@ -28,6 +28,7 @@ set(TEST_SRC if(UNIX) set(TEST_SRC ${TEST_SRC} + hipIpcOpenMemHandle.cc hipIpcGetMemHandle.cc hipIpcCloseMemHandle.cc ) diff --git a/tests/catch/unit/device/hipIpcOpenMemHandle.cc b/tests/catch/unit/device/hipIpcOpenMemHandle.cc new file mode 100644 index 0000000000..4edc7cd616 --- /dev/null +++ b/tests/catch/unit/device/hipIpcOpenMemHandle.cc @@ -0,0 +1,88 @@ +/* +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 +#include +#include + +#include +#include + +TEST_CASE("Unit_hipIpcOpenMemHandle_Negative_Open_In_Creating_Process") { + hipDeviceptr_t ptr1, ptr2; + hipIpcMemHandle_t handle; + HIP_CHECK(hipMalloc(reinterpret_cast(&ptr1), 1024)); + HIP_CHECK(hipIpcGetMemHandle(&handle, reinterpret_cast(ptr1))); + HIP_CHECK_ERROR( + hipIpcOpenMemHandle(reinterpret_cast(&ptr2), handle, hipIpcMemLazyEnablePeerAccess), + hipErrorInvalidContext); + HIP_CHECK(hipFree(reinterpret_cast(ptr1))); +} + +TEST_CASE("Unit_hipIpcOpenMemHandle_Negative_Open_In_Two_Contexts_Same_Device") { + int fd[2]; + REQUIRE(pipe(fd) == 0); + + // The fork must be performed before the runtime is initialized(so before any API that implicitly + // initializes it). The pipe in conjunction with wait is then used to impose total ordering + // between parent and child process. Because total ordering is imposed regular CATCH assertions + // should be safe to use + auto pid = fork(); + REQUIRE(pid >= 0); + if (pid == 0) { // child + REQUIRE(close(fd[1]) == 0); + + hipIpcMemHandle_t handle; + REQUIRE(read(fd[0], &handle, sizeof(handle)) >= 0); + REQUIRE(close(fd[0]) == 0); + + hipDeviceptr_t ptr_child; + HIP_CHECK(hipIpcOpenMemHandle(reinterpret_cast(&ptr_child), handle, + hipIpcMemLazyEnablePeerAccess)); + + HIP_CHECK(hipInit(0)); + hipCtx_t ctx; + HIP_CHECK(hipCtxCreate(&ctx, 0, 0)); + + hipDeviceptr_t ptr_child_ctx; + HIP_CHECK_ERROR(hipIpcOpenMemHandle(reinterpret_cast(&ptr_child_ctx), handle, + hipIpcMemLazyEnablePeerAccess), + hipErrorInvalidResourceHandle); + + exit(0); + } else { // parent + REQUIRE(close(fd[0]) == 0); + + hipDeviceptr_t ptr; + hipIpcMemHandle_t handle; + HIP_CHECK(hipMalloc(reinterpret_cast(&ptr), 1024)); + HIP_CHECK(hipIpcGetMemHandle(&handle, reinterpret_cast(ptr))); + + REQUIRE(write(fd[1], &handle, sizeof(handle)) >= 0); + REQUIRE(close(fd[1]) == 0); + + REQUIRE(wait(NULL) >= 0); + + HIP_CHECK(hipFree(reinterpret_cast(ptr))); + } +} \ No newline at end of file