Skip to content

Commit a1a45a9

Browse files
music-dinomusic-dino
committed
Merge remote-tracking branch 'upstream/develop' into hipMemAdvise_tests
2 parents c4def8e + cd4a634 commit a1a45a9

File tree

5 files changed

+202
-1
lines changed

5 files changed

+202
-1
lines changed

‎tests/catch/hipTestMain/config/config_amd_linux_common.json‎

Lines changed: 2 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -3,6 +3,8 @@
33
[
44
"Unit_hipStreamPerThread_DeviceReset_1",
55
"Unit_hipMallocManaged_OverSubscription",
6+
"Unit_hipDeviceGetSharedMemConfig_Positive_Basic",
7+
"Unit_hipDeviceGetSharedMemConfig_Positive_Threaded",
68
"Unit_hipDeviceGetCacheConfig_Positive_Basic",
79
"Unit_hipDeviceGetCacheConfig_Positive_Threaded",
810
"Unit_hipGetDeviceFlags_Positive_Context",

‎tests/catch/hipTestMain/config/config_amd_windows_common.json‎

Lines changed: 2 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -101,6 +101,8 @@
101101
"Unit_hipGetDeviceFlags_Positive_Context",
102102
"Unit_hipIpcCloseMemHandle_Negative_Close_In_Originating_Process",
103103
"Unit_hipDeviceGetPCIBusId_Negative_PartialFill",
104+
"Unit_hipDeviceGetSharedMemConfig_Positive_Basic",
105+
"Unit_hipDeviceGetSharedMemConfig_Positive_Threaded",
104106
"Unit_hipMemAdvise_AccessedBy_All_Devices",
105107
"Unit_hipMemAdvise_No_Flag_Interference"
106108
]

‎tests/catch/unit/device/CMakeLists.txt‎

Lines changed: 4 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -22,12 +22,15 @@ set(TEST_SRC
2222
hipDeviceEnableDisablePeerAccess.cc
2323
hipExtGetLinkTypeAndHopCount.cc
2424
hipDeviceSetLimit.cc
25+
hipDeviceSetGetSharedMemConfig.cc
2526
hipDeviceSetGetMemPool.cc
2627
)
2728

2829
if(UNIX)
2930
set(TEST_SRC ${TEST_SRC}
30-
hipIpcCloseMemHandle.cc)
31+
hipIpcGetMemHandle.cc
32+
hipIpcCloseMemHandle.cc
33+
)
3134
endif()
3235

3336
set_source_files_properties(hipGetDeviceCount.cc PROPERTIES COMPILE_FLAGS -std=c++17)

‎tests/catch/unit/device/hipDeviceSetGetSharedMemConfig.cc‎

Lines changed: 121 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,121 @@
1+
/*
2+
Copyright (c) 2022 Advanced Micro Devices, Inc. All rights reserved.
3+
Permission is hereby granted, free of charge, to any person obtaining a copy
4+
of this software and associated documentation files (the "Software"), to deal
5+
in the Software without restriction, including without limitation the rights
6+
to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
7+
copies of the Software, and to permit persons to whom the Software is
8+
furnished to do so, subject to the following conditions:
9+
10+
The above copyright notice and this permission notice shall be included in
11+
all copies or substantial portions of the Software.
12+
13+
THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
14+
IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
15+
FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
16+
AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
17+
LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
18+
OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN
19+
THE SOFTWARE.
20+
*/
21+
22+
#include <array>
23+
24+
#include <hip_test_common.hh>
25+
#include <threaded_zig_zag_test.hh>
26+
27+
namespace {
28+
constexpr std::array<hipSharedMemConfig, 3> kMemConfigs{
29+
hipSharedMemBankSizeDefault, hipSharedMemBankSizeFourByte, hipSharedMemBankSizeEightByte};
30+
} // anonymous namespace
31+
32+
TEST_CASE("Unit_hipDeviceSetSharedMemConfig_Positive_Basic") {
33+
const auto device = GENERATE(range(0, HipTest::getDeviceCount()));
34+
const auto mem_config = GENERATE(from_range(std::begin(kMemConfigs), std::end(kMemConfigs)));
35+
HIP_CHECK(hipSetDevice(device));
36+
INFO("Current device is " << device);
37+
38+
#if HT_AMD
39+
HIP_CHECK_ERROR(hipDeviceSetSharedMemConfig(mem_config), hipErrorNotSupported);
40+
#elif HT_NVIDIA
41+
HIP_CHECK(hipDeviceSetSharedMemConfig(mem_config));
42+
#endif
43+
}
44+
45+
TEST_CASE("Unit_hipDeviceSetSharedMemConfig_Negative_Parameters") {
46+
#if HT_AMD
47+
HIP_CHECK_ERROR(hipDeviceSetSharedMemConfig(static_cast<hipSharedMemConfig>(-1)),
48+
hipErrorNotSupported);
49+
#elif HT_NVIDIA
50+
HIP_CHECK_ERROR(hipDeviceSetSharedMemConfig(static_cast<hipSharedMemConfig>(-1)),
51+
hipErrorInvalidValue);
52+
#endif
53+
}
54+
55+
TEST_CASE("Unit_hipDeviceGetSharedMemConfig_Positive_Default") {
56+
const auto device = GENERATE(range(0, HipTest::getDeviceCount()));
57+
HIP_CHECK(hipSetDevice(device));
58+
INFO("Current device is " << device);
59+
60+
hipSharedMemConfig mem_config;
61+
HIP_CHECK(hipDeviceGetSharedMemConfig(&mem_config));
62+
REQUIRE(mem_config == hipSharedMemBankSizeFourByte);
63+
}
64+
65+
TEST_CASE("Unit_hipDeviceGetSharedMemConfig_Positive_Basic") {
66+
const auto device = GENERATE(range(0, HipTest::getDeviceCount()));
67+
const auto mem_config = GENERATE(from_range(std::begin(kMemConfigs), std::end(kMemConfigs)));
68+
HIP_CHECK(hipSetDevice(device));
69+
INFO("Current device is " << device);
70+
71+
HIP_CHECK(hipDeviceSetSharedMemConfig(mem_config));
72+
73+
hipSharedMemConfig returned_mem_config;
74+
HIP_CHECK(hipDeviceGetSharedMemConfig(&returned_mem_config));
75+
76+
int major = -1, minor = -1;
77+
HIP_CHECK(hipDeviceComputeCapability(&major, &minor, device));
78+
REQUIRE(major > 0);
79+
if (major == 3 /*Kepler*/) {
80+
REQUIRE(returned_mem_config == mem_config);
81+
} else {
82+
REQUIRE(returned_mem_config == hipSharedMemBankSizeFourByte);
83+
}
84+
}
85+
86+
TEST_CASE("Unit_hipDeviceGetSharedMemConfig_Positive_Threaded") {
87+
class HipDeviceGetSharedMemConfigTest
88+
: public ThreadedZigZagTest<HipDeviceGetSharedMemConfigTest> {
89+
public:
90+
HipDeviceGetSharedMemConfigTest(const hipSharedMemConfig mem_config)
91+
: mem_config_{mem_config} {}
92+
93+
void TestPart2() { HIP_CHECK_THREAD(hipDeviceSetSharedMemConfig(mem_config_)); }
94+
95+
void TestPart3() {
96+
hipSharedMemConfig returned_mem_config;
97+
HIP_CHECK(hipDeviceGetSharedMemConfig(&returned_mem_config));
98+
99+
int major = -1, minor = -1;
100+
HIP_CHECK(hipDeviceComputeCapability(&major, &minor, 0));
101+
REQUIRE(major > 0);
102+
if (major == 3 /*Kepler*/) {
103+
REQUIRE(returned_mem_config == mem_config_);
104+
} else {
105+
REQUIRE(returned_mem_config == hipSharedMemBankSizeFourByte);
106+
}
107+
}
108+
109+
private:
110+
const hipSharedMemConfig mem_config_;
111+
};
112+
113+
const auto mem_config = GENERATE(from_range(std::begin(kMemConfigs), std::end(kMemConfigs)));
114+
115+
HipDeviceGetSharedMemConfigTest test(mem_config);
116+
test.run();
117+
}
118+
119+
TEST_CASE("Unit_hipDeviceGetSharedMemConfig_Negative_Parameters") {
120+
HIP_CHECK_ERROR(hipDeviceGetSharedMemConfig(nullptr), hipErrorInvalidValue);
121+
}

‎tests/catch/unit/device/hipIpcGetMemHandle.cc‎

Lines changed: 73 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,73 @@
1+
/*
2+
Copyright (c) 2022 Advanced Micro Devices, Inc. All rights reserved.
3+
4+
Permission is hereby granted, free of charge, to any person obtaining a copy
5+
of this software and associated documentation files (the "Software"), to deal
6+
in the Software without restriction, including without limitation the rights
7+
to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
8+
copies of the Software, and to permit persons to whom the Software is
9+
furnished to do so, subject to the following conditions:
10+
11+
The above copyright notice and this permission notice shall be included in
12+
all copies or substantial portions of the Software.
13+
14+
THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
15+
IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
16+
FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
17+
AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
18+
LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
19+
OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN
20+
THE SOFTWARE.
21+
*/
22+
23+
#include <cstring>
24+
25+
#include <hip_test_common.hh>
26+
#include <hip/hip_runtime_api.h>
27+
28+
TEST_CASE("Unit_hipIpcGetMemHandle_Positive_Unique_Handles_Separate_Allocations") {
29+
void *ptr1, *ptr2;
30+
hipIpcMemHandle_t handle1, handle2;
31+
HIP_CHECK(hipMalloc(&ptr1, 1024));
32+
HIP_CHECK(hipMalloc(&ptr2, 1024));
33+
HIP_CHECK(hipIpcGetMemHandle(&handle1, ptr1));
34+
HIP_CHECK(hipIpcGetMemHandle(&handle2, ptr2));
35+
36+
CHECK(memcmp(&handle1, &handle2, sizeof(handle1)) != 0);
37+
38+
HIP_CHECK(hipFree(ptr1));
39+
HIP_CHECK(hipFree(ptr2));
40+
}
41+
42+
TEST_CASE("Unit_hipIpcGetMemHandle_Positive_Unique_Handles_Reused_Memory") {
43+
void *ptr1 = nullptr, *ptr2 = nullptr;
44+
hipIpcMemHandle_t handle1, handle2;
45+
HIP_CHECK(hipMalloc(&ptr1, 1024));
46+
HIP_CHECK(hipIpcGetMemHandle(&handle1, ptr1));
47+
HIP_CHECK(hipFree(ptr1));
48+
49+
HIP_CHECK(hipMalloc(&ptr2, 1024));
50+
HIP_CHECK(hipIpcGetMemHandle(&handle2, ptr2));
51+
52+
if (ptr1 == ptr2) CHECK(memcmp(&handle1, &handle2, sizeof(handle1)) != 0);
53+
54+
HIP_CHECK(hipFree(ptr2));
55+
}
56+
57+
TEST_CASE("Unit_hipIpcGetMemHandle_Negative_Handle_For_Freed_Memory") {
58+
void* ptr;
59+
hipIpcMemHandle_t handle;
60+
HIP_CHECK(hipMalloc(&ptr, 1024));
61+
HIP_CHECK(hipFree(ptr));
62+
HIP_CHECK_ERROR(hipIpcGetMemHandle(&handle, ptr), hipErrorInvalidValue);
63+
}
64+
65+
TEST_CASE("Unit_hipIpcGetMemHandle_Negative_Out_Of_Bound_Pointer") {
66+
int* ptr;
67+
constexpr size_t n = 1024;
68+
hipIpcMemHandle_t handle;
69+
HIP_CHECK(hipMalloc(reinterpret_cast<void**>(&ptr), n * sizeof(*ptr)));
70+
HIP_CHECK_ERROR(hipIpcGetMemHandle(&handle, reinterpret_cast<void*>(ptr + n)),
71+
hipErrorInvalidValue);
72+
HIP_CHECK(hipFree(reinterpret_cast<void*>(ptr)));
73+
}

0 commit comments

Comments
 (0)