AI/rocm-systems
2
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

SWDEV-299127 - Merge 'develop' into 'amd-staging'

Browse Source 
Change-Id: Ie193993a41286d5ec789abebff53ee2ed9130930
This commit is contained in:
Jenkins
2022-11-08 19:10:03 -05:00
parent 83fb425385 809149ecc8
commit d44e4d00ba
10 changed files with 307 additions and 39 deletions
Download Patch File Download Diff File Expand all files Collapse all files
tests/catch/CMakeLists.txt
+5 -2
View File
@@ -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})
tests/catch/hipTestMain/config/config_amd_linux_common.json
+3
View File
@@ -3,10 +3,13 @@
[
"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",
"Unit_hipIpcCloseMemHandle_Negative_Close_In_Originating_Process",
"Unit_hipIpcOpenMemHandle_Negative_Open_In_Creating_Process",
"Unit_hipDeviceGetPCIBusId_Negative_PartialFill"
]
}
tests/catch/hipTestMain/config/config_amd_windows_common.json
+4 -25
View File
@@ -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",
@@ -100,6 +76,9 @@
"Unit_hipStreamValue_Wait64_Blocking_NoMask_Nor",
"Unit_hipGetDeviceFlags_Positive_Context",
"Unit_hipIpcCloseMemHandle_Negative_Close_In_Originating_Process",
"Unit_hipDeviceGetPCIBusId_Negative_PartialFill"
"Unit_hipIpcOpenMemHandle_Negative_Open_In_Creating_Process",
"Unit_hipDeviceGetPCIBusId_Negative_PartialFill",
"Unit_hipDeviceGetSharedMemConfig_Positive_Basic",
"Unit_hipDeviceGetSharedMemConfig_Positive_Threaded"
]
}
tests/catch/packaging/hip-tests.txt
+7
View File
@@ -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)
tests/catch/unit/device/CMakeLists.txt
+5 -1
View File
@@ -22,12 +22,16 @@ set(TEST_SRC
hipDeviceEnableDisablePeerAccess.cc
hipExtGetLinkTypeAndHopCount.cc
hipDeviceSetLimit.cc
hipDeviceSetGetSharedMemConfig.cc
hipDeviceSetGetMemPool.cc
)
if(UNIX)
set(TEST_SRC ${TEST_SRC}
hipIpcCloseMemHandle.cc)
hipIpcOpenMemHandle.cc
hipIpcGetMemHandle.cc
hipIpcCloseMemHandle.cc
)
endif()
set_source_files_properties(hipGetDeviceCount.cc PROPERTIES COMPILE_FLAGS -std=c++17)
tests/catch/unit/device/hipDeviceSetGetSharedMemConfig.cc
+121
View File
@@ -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 <array>
#include <hip_test_common.hh>
#include <threaded_zig_zag_test.hh>
namespace {
constexpr std::array<hipSharedMemConfig, 3> 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<hipSharedMemConfig>(-1)),
hipErrorNotSupported);
#elif HT_NVIDIA
HIP_CHECK_ERROR(hipDeviceSetSharedMemConfig(static_cast<hipSharedMemConfig>(-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<HipDeviceGetSharedMemConfigTest> {
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);
}
tests/catch/unit/device/hipIpcGetMemHandle.cc
+73
View File
@@ -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 <cstring>
#include <hip_test_common.hh>
#include <hip/hip_runtime_api.h>
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<void**>(&ptr), n * sizeof(*ptr)));
HIP_CHECK_ERROR(hipIpcGetMemHandle(&handle, reinterpret_cast<void*>(ptr + n)),
hipErrorInvalidValue);
HIP_CHECK(hipFree(reinterpret_cast<void*>(ptr)));
}
tests/catch/unit/device/hipIpcOpenMemHandle.cc
+88
View File
@@ -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 <cstring>
#include <unistd.h>
#include <sys/types.h>
#include <sys/wait.h>
#include <hip_test_common.hh>
#include <hip/hip_runtime_api.h>
TEST_CASE("Unit_hipIpcOpenMemHandle_Negative_Open_In_Creating_Process") {
hipDeviceptr_t ptr1, ptr2;
hipIpcMemHandle_t handle;
HIP_CHECK(hipMalloc(reinterpret_cast<void**>(&ptr1), 1024));
HIP_CHECK(hipIpcGetMemHandle(&handle, reinterpret_cast<void*>(ptr1)));
HIP_CHECK_ERROR(
hipIpcOpenMemHandle(reinterpret_cast<void**>(&ptr2), handle, hipIpcMemLazyEnablePeerAccess),
hipErrorInvalidContext);
HIP_CHECK(hipFree(reinterpret_cast<void*>(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<void**>(&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<void**>(&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<void**>(&ptr), 1024));
HIP_CHECK(hipIpcGetMemHandle(&handle, reinterpret_cast<void*>(ptr)));
REQUIRE(write(fd[1], &handle, sizeof(handle)) >= 0);
REQUIRE(close(fd[1]) == 0);
REQUIRE(wait(NULL) >= 0);
HIP_CHECK(hipFree(reinterpret_cast<void*>(ptr)));
}
}
tests/catch/unit/event/hipEventDestroy.cc
-5
View File
@@ -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);
tests/catch/unit/stream/hipStreamDestroy.cc
+1 -6
View File
@@ -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;