From f729e892a27645a289ef4932dfea38f2fdbe1b8a Mon Sep 17 00:00:00 2001 From: mbhiutra Date: Wed, 25 Oct 2023 16:24:48 +0530 Subject: [PATCH] SWDEV-422641 - [catch2][dtest] Added test for Annoying truncation of 'size_t' to 'unsigned int' inside hipMemcpy3D Change-Id: I0faef0278acbb82186839fcd556af508b6fc84ec --- catch/unit/memory/hipDrvMemcpy3DAsync_old.cc | 154 ++++++++++-- catch/unit/memory/hipDrvMemcpy3D_old.cc | 154 ++++++++++-- catch/unit/memory/hipMemcpy2D.cc | 251 ++++++++++++++----- catch/unit/memory/hipMemcpy2DAsync.cc | 247 ++++++++++++++---- catch/unit/memory/hipMemcpy3DAsync_old.cc | 130 +++++++--- catch/unit/memory/hipMemcpy3D_old.cc | 143 +++++++++-- 6 files changed, 873 insertions(+), 206 deletions(-) diff --git a/catch/unit/memory/hipDrvMemcpy3DAsync_old.cc b/catch/unit/memory/hipDrvMemcpy3DAsync_old.cc index e556e52997..c44992746c 100644 --- a/catch/unit/memory/hipDrvMemcpy3DAsync_old.cc +++ b/catch/unit/memory/hipDrvMemcpy3DAsync_old.cc @@ -1,5 +1,5 @@ /* -Copyright (c) 2022 Advanced Micro Devices, Inc. All rights reserved. +Copyright (c) 2022-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 @@ -16,6 +16,15 @@ 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. */ + +/** + * @addtogroup hipDrvMemcpy3DAsync hipDrvMemcpy3DAsync + * @{ + * @ingroup MemoryTest + * `hipMemcpy3DAsync(const hipMemcpy3DParms* p, hipStream_t stream)` - + * Copies data between 3D objects. + */ + /* * Test Scenarios * 1. Verifying hipDrvMemcpy3DAsync API for H2A,A2A,A2H scenarios @@ -33,8 +42,8 @@ THE SOFTWARE. * Scenario 5&6 are excluded in CUDA platform */ -#include "hip_test_common.hh" -#include "hip_test_checkers.hh" +#include +#include template class DrvMemcpy3DAsync { @@ -448,10 +457,21 @@ void DrvMemcpy3DAsync::DeAllocateMemory() { free(hData); } -/* Verifying hipDrvMemcpy3DAsync API Host to Array for different datatypes */ -TEMPLATE_TEST_CASE("Unit_hipDrvMemcpy3DAsync_MultipleDataTypes", "", uint8_t, int, float) { - CHECK_IMAGE_SUPPORT +/** + * Test Description + * ------------------------ + * - Verifying hipDrvMemcpy3DAsync API Host to Array for different datatypes + * Test source + * ------------------------ + * - unit/memory/hipDrvMemcpy3DAsync_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ +TEMPLATE_TEST_CASE("Unit_hipDrvMemcpy3DAsync_MultipleDataTypes", "", + uint8_t, int, float) { + CHECK_IMAGE_SUPPORT for (int i = 1; i < 25; i++) { if (std::is_same::value) { DrvMemcpy3DAsync memcpy3d_float(i, i, i, @@ -469,15 +489,36 @@ TEMPLATE_TEST_CASE("Unit_hipDrvMemcpy3DAsync_MultipleDataTypes", "", uint8_t, in } } -/* This testcase verifies H2D copy of hipDrvMemcpy3DAsync API */ +/** + * Test Description + * ------------------------ + * - This testcase verifies H2D copy of hipDrvMemcpy3DAsync API + * Test source + * ------------------------ + * - unit/memory/hipDrvMemcpy3DAsync_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + TEST_CASE("Unit_hipDrvMemcpy3DAsync_HosttoDevice") { CHECK_IMAGE_SUPPORT - DrvMemcpy3DAsync memcpy3d_D2H_float(10, 10, 1, HIP_AD_FORMAT_FLOAT); memcpy3d_D2H_float.HostDevice_DrvMemcpy3DAsync(); } -/* This testcase verifies negative scenarios of hipDrvMemcpy3DAsync API */ +/** + * Test Description + * ------------------------ + * - This testcase verifies negative scenarios of hipDrvMemcpy3DAsync API + * Test source + * ------------------------ + * - unit/memory/hipDrvMemcpy3DAsync_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + #if HT_NVIDIA TEST_CASE("Unit_hipDrvMemcpy3DAsync_Negative") { DrvMemcpy3DAsync memcpy3d(10, 10, 1, HIP_AD_FORMAT_FLOAT); @@ -485,21 +526,40 @@ TEST_CASE("Unit_hipDrvMemcpy3DAsync_Negative") { } #endif -/* This testcase verifies extent validation scenarios of - hipDrvMemcpy3DAsync API */ +/** + * Test Description + * ------------------------ + * - This testcase verifies extent validation scenarios of + hipDrvMemcpy3DAsync API + * Test source + * ------------------------ + * - unit/memory/hipDrvMemcpy3DAsync_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + TEST_CASE("Unit_hipDrvMemcpy3DAsync_ExtentValidation") { CHECK_IMAGE_SUPPORT - DrvMemcpy3DAsync memcpy3d(10, 10, 1, HIP_AD_FORMAT_FLOAT); memcpy3d.Extent_Validation(); } -/* This testcase verifies H2D copy in device context -change scenario for hipDrvMemcpy3DAsync API */ -#if HT_AMD +/** + * Test Description + * ------------------------ + * - This testcase verifies H2D copy in device context + change scenario for hipDrvMemcpy3DAsync API + * Test source + * ------------------------ + * - unit/memory/hipDrvMemcpy3DAsync_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + TEST_CASE("Unit_hipDrvMemcpy3DAsync_H2DDeviceContextChange") { CHECK_IMAGE_SUPPORT - int numDevices = 0; HIP_CHECK(hipGetDeviceCount(&numDevices)); if (numDevices > 1) { @@ -510,12 +570,21 @@ TEST_CASE("Unit_hipDrvMemcpy3DAsync_H2DDeviceContextChange") { } } +/** + * Test Description + * ------------------------ + * - This testcase verifies Host to Array copy in device context + change scenario for hipDrvMemcpy3DAsync API + * Test source + * ------------------------ + * - unit/memory/hipDrvMemcpy3DAsync_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ -/* This testcase verifies Host to Array copy in device context -change scenario for hipDrvMemcpy3DAsync API */ TEST_CASE("Unit_hipDrvMemcpy3DAsync_Host2ArrayDeviceContextChange") { CHECK_IMAGE_SUPPORT - int numDevices = 0; HIP_CHECK(hipGetDeviceCount(&numDevices)); if (numDevices > 1) { @@ -525,6 +594,51 @@ TEST_CASE("Unit_hipDrvMemcpy3DAsync_Host2ArrayDeviceContextChange") { SUCCEED("skipped testcase as Device count is < 2"); } } -#endif +/** + * Test Description + * ------------------------ + * - This testcase performs multidevice size check on DrvMemcpy3DAsync API + 1. Verify with 128 for all height, width & depth value + 2. Verify with 256 for height and 128 for width & depth value + 3. Verify with 256 for width and 128 for height & depth value + 4. Verify with 256 for depth and 128 for height & width value + * Test source + * ------------------------ + * - unit/memory/hipDrvMemcpy3DAsync_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ +TEST_CASE("Unit_hipDrvMemcpy3DAsync_multiDevice_Basic_Size_Test") { + CHECK_IMAGE_SUPPORT + constexpr int size_128b = 128, size_256b = 256; + int numDevices = 0; + HIP_CHECK(hipGetDeviceCount(&numDevices)); + + for (int i=0; i < numDevices; i++) { + HIP_CHECK(hipSetDevice(i)); + + SECTION("Verify with 128 for all height, width & depth value") { + DrvMemcpy3DAsync memcpy3d(size_128b, size_128b, size_128b, + HIP_AD_FORMAT_SIGNED_INT32); + memcpy3d.HostArray_DrvMemcpy3DAsync(); + } + SECTION("Verify with 256 for height and 128 for width & depth value") { + DrvMemcpy3DAsync memcpy3d(size_256b, size_128b, size_128b, + HIP_AD_FORMAT_SIGNED_INT32); + memcpy3d.HostArray_DrvMemcpy3DAsync(); + } + SECTION("Verify with 256 for width and 128 for height & depth value") { + DrvMemcpy3DAsync memcpy3d(size_128b, size_256b, size_128b, + HIP_AD_FORMAT_FLOAT); + memcpy3d.HostArray_DrvMemcpy3DAsync(); + } + SECTION("Verify with 256 for depth and 128 for height & width value") { + DrvMemcpy3DAsync memcpy3d(size_128b, size_128b, size_256b, + HIP_AD_FORMAT_SIGNED_INT32); + memcpy3d.HostArray_DrvMemcpy3DAsync(); + } + } +} diff --git a/catch/unit/memory/hipDrvMemcpy3D_old.cc b/catch/unit/memory/hipDrvMemcpy3D_old.cc index ef580d0315..0e39deb95f 100644 --- a/catch/unit/memory/hipDrvMemcpy3D_old.cc +++ b/catch/unit/memory/hipDrvMemcpy3D_old.cc @@ -1,5 +1,5 @@ /* -Copyright (c) 2022 Advanced Micro Devices, Inc. All rights reserved. +Copyright (c) 2022-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 @@ -16,6 +16,15 @@ 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. */ + +/** + * @addtogroup hipDrvMemcpy3D hipDrvMemcpy3D + * @{ + * @ingroup MemoryTest + * `hipMemcpy3D(const hipMemcpy3DParms* p)` - + * Copies data between 3D objects. + */ + /* * Test Scenarios * 1. Verifying hipDrvMemcpy3D API for H2A,A2A,A2H scenarios @@ -33,8 +42,8 @@ THE SOFTWARE. * Scenario 5&6 are not supported in CUDA platform */ -#include "hip_test_common.hh" -#include "hip_test_checkers.hh" +#include +#include template class DrvMemcpy3D { @@ -433,10 +442,21 @@ void DrvMemcpy3D::DeAllocateMemory() { free(hData); } -/* Verifying hipDrvMemcpy3D API Host to Array for different datatypes */ -TEMPLATE_TEST_CASE("Unit_hipDrvMemcpy3D_MultipleDataTypes", "", uint8_t, int, float) { - CHECK_IMAGE_SUPPORT +/** + * Test Description + * ------------------------ + * - Verifying hipDrvMemcpy3D API Host to Array for different datatypes + * Test source + * ------------------------ + * - unit/memory/hipDrvMemcpy3D_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ +TEMPLATE_TEST_CASE("Unit_hipDrvMemcpy3D_MultipleDataTypes", "", + uint8_t, int, float) { + CHECK_IMAGE_SUPPORT for (int i = 1; i < 25; i++) { if (std::is_same::value) { DrvMemcpy3D memcpy3d_float(i, i, i, HIP_AD_FORMAT_FLOAT); @@ -451,36 +471,76 @@ TEMPLATE_TEST_CASE("Unit_hipDrvMemcpy3D_MultipleDataTypes", "", uint8_t, int, fl } } -/* This testcase verifies H2D copy of hipDrvMemcpy3D API */ +/** + * Test Description + * ------------------------ + * - This testcase verifies H2D copy of hipDrvMemcpy3D API + * Test source + * ------------------------ + * - unit/memory/hipDrvMemcpy3D_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + TEST_CASE("Unit_hipDrvMemcpy3D_HosttoDevice") { CHECK_IMAGE_SUPPORT - DrvMemcpy3D memcpy3d_D2H_float(10, 10, 1, HIP_AD_FORMAT_FLOAT); memcpy3d_D2H_float.HostDevice_DrvMemcpy3D(); } -/* This testcase verifies negative scenarios of hipDrvMemcpy3D API */ #if HT_NVIDIA +/** + * Test Description + * ------------------------ + * - This testcase verifies negative scenarios of hipDrvMemcpy3D API + * Test source + * ------------------------ + * - unit/memory/hipDrvMemcpy3D_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + TEST_CASE("Unit_hipDrvMemcpy3D_Negative") { DrvMemcpy3D memcpy3d(10, 10, 1, HIP_AD_FORMAT_FLOAT); memcpy3d.NegativeTests(); } #endif -/* This testcase verifies extent validation scenarios of hipDrvMemcpy3D API */ +/** + * Test Description + * ------------------------ + * - This testcase verifies extent validation scenarios of hipDrvMemcpy3D API + * Test source + * ------------------------ + * - unit/memory/hipDrvMemcpy3D_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + TEST_CASE("Unit_hipDrvMemcpy3D_ExtentValidation") { CHECK_IMAGE_SUPPORT - DrvMemcpy3D memcpy3d(10, 10, 1, HIP_AD_FORMAT_FLOAT); memcpy3d.Extent_Validation(); } -#if HT_AMD -/* This testcase verifies H2D copy in device context -change scenario for hipDrvMemcpy3D API */ +/** + * Test Description + * ------------------------ + * - This testcase verifies H2D copy in device context + change scenario for hipDrvMemcpy3D API + * Test source + * ------------------------ + * - unit/memory/hipDrvMemcpy3D_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + TEST_CASE("Unit_hipDrvMemcpy3D_H2DDeviceContextChange") { CHECK_IMAGE_SUPPORT - int numDevices = 0; HIP_CHECK(hipGetDeviceCount(&numDevices)); if (numDevices > 1) { @@ -491,12 +551,21 @@ TEST_CASE("Unit_hipDrvMemcpy3D_H2DDeviceContextChange") { } } +/** + * Test Description + * ------------------------ + * - This testcase verifies Host to Array copy in device context + change scenario for hipDrvMemcpy3D API + * Test source + * ------------------------ + * - unit/memory/hipDrvMemcpy3D_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ -/* This testcase verifies Host to Array copy in device context -change scenario for hipDrvMemcpy3D API */ TEST_CASE("Unit_hipDrvMemcpy3D_Host2ArrayDeviceContextChange") { CHECK_IMAGE_SUPPORT - int numDevices = 0; HIP_CHECK(hipGetDeviceCount(&numDevices)); if (numDevices > 1) { @@ -506,4 +575,51 @@ TEST_CASE("Unit_hipDrvMemcpy3D_Host2ArrayDeviceContextChange") { SUCCEED("skipped testcase as Device count is < 2"); } } -#endif + +/** + * Test Description + * ------------------------ + * - This testcase performs multidevice size check on hipDrvMemcpy3D API + 1. Verify with 128 for all height, width & depth value + 2. Verify with 256 for height and 128 for width & depth value + 3. Verify with 256 for width and 128 for height & depth value + 4. Verify with 256 for depth and 128 for height & width value + * Test source + * ------------------------ + * - unit/memory/hipDrvMemcpy3D_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + +TEST_CASE("Unit_hipDrvMemcpy3D_multiDevice_Basic_Size_Test") { + CHECK_IMAGE_SUPPORT + constexpr int size_128b = 128, size_256b = 256; + int numDevices = 0; + HIP_CHECK(hipGetDeviceCount(&numDevices)); + + for (int i=0; i < numDevices; i++) { + HIP_CHECK(hipSetDevice(i)); + + SECTION("Verify with 128 for all height, width & depth value") { + DrvMemcpy3D memcpy3d(size_128b, size_128b, size_128b, + HIP_AD_FORMAT_SIGNED_INT32); + memcpy3d.HostArray_DrvMemcpy3D(); + } + SECTION("Verify with 256 for height and 128 for width & depth value") { + DrvMemcpy3D memcpy3d(size_256b, size_128b, size_128b, + HIP_AD_FORMAT_SIGNED_INT32); + memcpy3d.HostArray_DrvMemcpy3D(); + } + SECTION("Verify with 256 for width and 128 for height & depth value") { + DrvMemcpy3D memcpy3d(size_128b, size_256b, size_128b, + HIP_AD_FORMAT_FLOAT); + memcpy3d.HostArray_DrvMemcpy3D(); + } + SECTION("Verify with 256 for depth and 128 for height & width value") { + DrvMemcpy3D memcpy3d(size_128b, size_128b, size_256b, + HIP_AD_FORMAT_SIGNED_INT32); + memcpy3d.HostArray_DrvMemcpy3D(); + } + } +} diff --git a/catch/unit/memory/hipMemcpy2D.cc b/catch/unit/memory/hipMemcpy2D.cc index 8820a2b3fa..d7e21e42f2 100644 --- a/catch/unit/memory/hipMemcpy2D.cc +++ b/catch/unit/memory/hipMemcpy2D.cc @@ -1,5 +1,5 @@ /* -Copyright (c) 2021 Advanced Micro Devices, Inc. All rights reserved. +Copyright (c) 2021-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 @@ -17,6 +17,16 @@ OUT OF OR INN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE SOFTWARE. */ +/** + * @addtogroup hipMemcpy2D hipMemcpy2D + * @{ + * @ingroup MemcpyTest + * `hipMemcpy2D(void* dst, size_t dpitch, const void* src, + * size_t spitch, size_t width, size_t height, + * hipMemcpyKind kind)` - + * Copies data between host and device. + */ + // Testcase Description: // 1) Verifies the working of Memcpy2D API negative scenarios by // Pass NULL to destination pointer @@ -43,21 +53,30 @@ static constexpr auto NUM_H{16}; static constexpr auto COLUMNS{8}; static constexpr auto ROWS{8}; -/* -This testcases performs the following scenarios of hipMemcpy2D API on same GPU -1. H2D-D2D-D2H for Host Memory<-->Device Memory -2. H2D-D2D-D2H for Pinned Host Memory<-->Device Memory +/** + * Test Description + * ------------------------ + * - This testcases performs the following scenarios of hipMemcpy2D API on same GPU + 1. H2D-D2D-D2H for Host Memory<-->Device Memory + 2. H2D-D2D-D2H for Pinned Host Memory<-->Device Memory -Input : "A_h" initialized based on data type - "A_h" --> "A_d" using H2D copy - "A_d" --> "B_d" using D2D copy - "B_d" --> "B_h" using D2H copy -Output: Validating A_h with B_h both should be equal for - the number of COLUMNS and ROWS copied -*/ -TEMPLATE_TEST_CASE("Unit_hipMemcpy2D_H2D-D2D-D2H", "", int, float, double) { + Input : "A_h" initialized based on data type + "A_h" --> "A_d" using H2D copy + "A_d" --> "B_d" using D2D copy + "B_d" --> "B_h" using D2H copy + Output: Validating A_h with B_h both should be equal for + the number of COLUMNS and ROWS copied + * Test source + * ------------------------ + * - unit/memory/hipMemcpy2D.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + +TEMPLATE_TEST_CASE("Unit_hipMemcpy2D_H2D-D2D-D2H", "" + , int, float, double) { CHECK_IMAGE_SUPPORT - // 1 refers to pinned host memory auto mem_type = GENERATE(0, 1); HIP_CHECK(hipSetDevice(0)); @@ -84,7 +103,8 @@ TEMPLATE_TEST_CASE("Unit_hipMemcpy2D_H2D-D2D-D2H", "", int, float, double) { // Host to Device HIP_CHECK(hipMemcpy2D(A_d, pitch_A, A_h, COLUMNS*sizeof(TestType), - COLUMNS*sizeof(TestType), ROWS, hipMemcpyHostToDevice)); + COLUMNS*sizeof(TestType), ROWS, + hipMemcpyHostToDevice)); // Performs D2D on same GPU device HIP_CHECK(hipMemcpy2D(B_d, pitch_B, A_d, @@ -99,7 +119,6 @@ TEMPLATE_TEST_CASE("Unit_hipMemcpy2D_H2D-D2D-D2H", "", int, float, double) { // Validating the result REQUIRE(HipTest::checkArray(A_h, B_h, COLUMNS, ROWS) == true); - // DeAllocating the memory HIP_CHECK(hipFree(A_d)); HIP_CHECK(hipFree(B_d)); @@ -111,23 +130,33 @@ TEMPLATE_TEST_CASE("Unit_hipMemcpy2D_H2D-D2D-D2H", "", int, float, double) { A_h, B_h, C_h, false); } } -/* -This testcase performs the following scenarios of hipMemcpy2D API on same GPU. -1. H2D-D2D-D2H for Host Memory<-->Device Memory -2. H2D-D2D-D2H for Pinned Host Memory<-->Device Memory -The src and dst input pointers to hipMemCpy2D add an offset to the pointers -returned by the allocation functions. -Input : "A_h" initialized based on data type - "A_h" --> "A_d" using H2D copy - "A_d" --> "B_d" using D2D copy - "B_d" --> "B_h" using D2H copy -Output: Validating A_h with B_h both should be equal for - the number of COLUMNS and ROWS copied -*/ -TEMPLATE_TEST_CASE("Unit_hipMemcpy2D_H2D-D2D-D2H_WithOffset", "", int, float, double) { +/** + * Test Description + * ------------------------ + * - This testcase performs the following scenarios of hipMemcpy2D API on same GPU. + 1. H2D-D2D-D2H for Host Memory<-->Device Memory + 2. H2D-D2D-D2H for Pinned Host Memory<-->Device Memory + The src and dst input pointers to hipMemCpy2D add an offset to the pointers + returned by the allocation functions. + + Input : "A_h" initialized based on data type + "A_h" --> "A_d" using H2D copy + "A_d" --> "B_d" using D2D copy + "B_d" --> "B_h" using D2H copy + Output: Validating A_h with B_h both should be equal for + the number of COLUMNS and ROWS copied + * Test source + * ------------------------ + * - unit/memory/hipMemcpy2D.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + +TEMPLATE_TEST_CASE("Unit_hipMemcpy2D_H2D-D2D-D2H_WithOffset", "" + , int, float, double) { CHECK_IMAGE_SUPPORT - // 1 refers to pinned host memory auto mem_type = GENERATE(0, 1); HIP_CHECK(hipSetDevice(0)); @@ -153,16 +182,19 @@ TEMPLATE_TEST_CASE("Unit_hipMemcpy2D_H2D-D2D-D2H_WithOffset", "", int, float, do HipTest::setDefaultData(NUM_W*NUM_H, A_h, B_h, C_h); // Host to Device - HIP_CHECK(hipMemcpy2D(A_d+COLUMNS*sizeof(TestType), pitch_A, A_h, COLUMNS*sizeof(TestType), - COLUMNS*sizeof(TestType), ROWS, hipMemcpyHostToDevice)); + HIP_CHECK(hipMemcpy2D(A_d+COLUMNS*sizeof(TestType), pitch_A, A_h, + COLUMNS*sizeof(TestType), COLUMNS*sizeof(TestType), + ROWS, hipMemcpyHostToDevice)); // Performs D2D on same GPU device - HIP_CHECK(hipMemcpy2D(B_d+COLUMNS*sizeof(TestType), pitch_B, A_d+COLUMNS*sizeof(TestType), + HIP_CHECK(hipMemcpy2D(B_d+COLUMNS*sizeof(TestType), pitch_B, + A_d+COLUMNS*sizeof(TestType), pitch_A, COLUMNS*sizeof(TestType), ROWS, hipMemcpyDeviceToDevice)); // hipMemcpy2D Device to Host - HIP_CHECK(hipMemcpy2D(B_h, COLUMNS*sizeof(TestType), B_d+COLUMNS*sizeof(TestType), pitch_B, + HIP_CHECK(hipMemcpy2D(B_h, COLUMNS*sizeof(TestType), + B_d+COLUMNS*sizeof(TestType), pitch_B, COLUMNS*sizeof(TestType), ROWS, hipMemcpyDeviceToHost)); @@ -183,23 +215,32 @@ TEMPLATE_TEST_CASE("Unit_hipMemcpy2D_H2D-D2D-D2H_WithOffset", "", int, float, do } } -/* -This testcases performs the following scenarios of hipMemcpy2D API on Peer GPU -1. H2D-D2D-D2H for Host Memory<-->Device Memory -2. H2D-D2D-D2H for Pinned Host Memory<-->Device Memory -3. Device context change where memory is allocated in GPU-0 - and API is trigerred from GPU-1 +/** + * Test Description + * ------------------------ + * - This testcases performs the following scenarios of hipMemcpy2D API on Peer GPU + 1. H2D-D2D-D2H for Host Memory<-->Device Memory + 2. H2D-D2D-D2H for Pinned Host Memory<-->Device Memory + 3. Device context change where memory is allocated in GPU-0 + and API is trigerred from GPU-1 -Input : "A_h" initialized based on data type - "A_h" --> "A_d" using H2D copy - "A_d" --> "X_d" using D2D copy - "X_d" --> "B_h" using D2H copy -Output: Validating A_h with B_h both should be equal for - the number of COLUMNS and ROWS copied -*/ -TEMPLATE_TEST_CASE("Unit_hipMemcpy2D_multiDevice-D2D", "", int, float, double) { + Input : "A_h" initialized based on data type + "A_h" --> "A_d" using H2D copy + "A_d" --> "X_d" using D2D copy + "X_d" --> "B_h" using D2H copy + Output: Validating A_h with B_h both should be equal for + the number of COLUMNS and ROWS copied + * Test source + * ------------------------ + * - unit/memory/hipMemcpy2D.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + +TEMPLATE_TEST_CASE("Unit_hipMemcpy2D_multiDevice-D2D", "" + , int, float, double) { CHECK_IMAGE_SUPPORT - auto mem_type = GENERATE(0, 1); int numDevices = 0; int canAccessPeer = 0; @@ -268,12 +309,20 @@ TEMPLATE_TEST_CASE("Unit_hipMemcpy2D_multiDevice-D2D", "", int, float, double) { } } -/* -This Testcase verifies the null size checks of hipMemcpy2D API -*/ +/** + * Test Description + * ------------------------ + * - This Testcase verifies the null size checks of hipMemcpy2D API + * Test source + * ------------------------ + * - unit/memory/hipMemcpy2D.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + TEST_CASE("Unit_hipMemcpy2D_SizeCheck") { CHECK_IMAGE_SUPPORT - HIP_CHECK(hipSetDevice(0)); int* A_h{nullptr}, *A_d{nullptr}; size_t pitch_A; @@ -317,12 +366,20 @@ TEST_CASE("Unit_hipMemcpy2D_SizeCheck") { free(A_h); } -/* -This Testcase verifies all the negative scenarios of hipMemcpy2D API -*/ +/** + * Test Description + * ------------------------ + * - This Testcase verifies all the negative scenarios of hipMemcpy2D API + * Test source + * ------------------------ + * - unit/memory/hipMemcpy2D.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + TEST_CASE("Unit_hipMemcpy2D_Negative") { CHECK_IMAGE_SUPPORT - HIP_CHECK(hipSetDevice(0)); int* A_h{nullptr}, *A_d{nullptr}; size_t pitch_A; @@ -359,3 +416,81 @@ TEST_CASE("Unit_hipMemcpy2D_Negative") { HIP_CHECK(hipFree(A_d)); free(A_h); } + +static void hipMemcpy2D_Basic_Size_Test(size_t inc) { + constexpr int defaultProgramSize = 256 * 1024 * 1024; + constexpr int N = 2; + constexpr int value = 42; + int *in, *out, *dev; + size_t newSize = 0, inp = 0; + size_t size = sizeof(int) * N * inc; + + size_t free, total; + HIP_CHECK(hipMemGetInfo(&free, &total)); + + if ( free < 2 * size ) + newSize = ( free - defaultProgramSize ) / 2; + else + newSize = size; + + INFO("Array size: " << size/1024.0/1024.0 << " MB or " << size << " Bytes."); + INFO("Free memory: " << free/1024.0/1024.0 << " MB or " << free << " Bytes"); + INFO("NewSize:" << newSize/1024.0/1024.0 << "MB or " << newSize << " Bytes"); + + HIP_CHECK(hipHostMalloc(&in, newSize)); + HIP_CHECK(hipHostMalloc(&out, newSize)); + HIP_CHECK(hipMalloc(&dev, newSize)); + + inp = newSize / (sizeof(int) * N); + for (size_t i=0; i < N; i++) { + in[i * inp] = value; + } + + size_t pitch = sizeof(int) * inp; + + HIP_CHECK(hipMemcpy2D(dev, pitch, in, pitch, sizeof(int), + N, hipMemcpyHostToDevice)); + HIP_CHECK(hipMemcpy2D(out, pitch, dev, pitch, sizeof(int), + N, hipMemcpyDeviceToHost)); + + for (size_t i=0; i < N; i++) { + REQUIRE(out[i * inp] == value); + } + + HIP_CHECK(hipFree(dev)); + HIP_CHECK(hipHostFree(in)); + HIP_CHECK(hipHostFree(out)); +} + +/** + * Test Description + * ------------------------ + * - This testcase performs multidevice size check on hipMemcpy2D API + 1. Verify hipMemcpy2D with 1 << 20 size + 2. Verify hipMemcpy2D with 1 << 21 size + * Test source + * ------------------------ + * - unit/memory/hipMemcpy2D.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + +TEST_CASE("Unit_hipMemcpy2D_multiDevice_Basic_Size_Test") { + CHECK_IMAGE_SUPPORT + size_t input = 1 << 20; + int numDevices = 0; + HIP_CHECK(hipGetDeviceCount(&numDevices)); + + for (int i=0; i < numDevices; i++) { + HIP_CHECK(hipSetDevice(i)); + + SECTION("Verify hipMemcpy2D with 1 << 20 size") { + hipMemcpy2D_Basic_Size_Test(input); + } + SECTION("Verify hipMemcpy2D with 1 << 21 size") { + input <<= 1; + hipMemcpy2D_Basic_Size_Test(input); + } + } +} diff --git a/catch/unit/memory/hipMemcpy2DAsync.cc b/catch/unit/memory/hipMemcpy2DAsync.cc index 645d9ddc4e..1ca39bd6c9 100644 --- a/catch/unit/memory/hipMemcpy2DAsync.cc +++ b/catch/unit/memory/hipMemcpy2DAsync.cc @@ -1,5 +1,5 @@ /* -Copyright (c) 2021 Advanced Micro Devices, Inc. All rights reserved. +Copyright (c) 2021-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 @@ -17,6 +17,16 @@ OUT OF OR INN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE SOFTWARE. */ +/** + * @addtogroup hipMemcpy2DAsync hipMemcpy2DAsync + * @{ + * @ingroup MemcpyTest + * `hipMemcpy2DAsync(void* dst, size_t dpitch, const void* src, + * size_t spitch, size_t width, size_t height, + * hipMemcpyKind kind, hipStream_t stream = 0 )` - + * Copies data between host and device. + */ + // Testcase Description: // 1) Verifies the working of Memcpy2DAsync API negative scenarios by // Pass NULL to destination pointer @@ -44,21 +54,30 @@ static constexpr auto NUM_H{16}; static constexpr auto COLUMNS{6}; static constexpr auto ROWS{6}; -/* -This performs the following scenarios of hipMemcpy2DAsync API on same GPU -1. H2D-D2D-D2H for Host Memory<-->Device Memory -2. H2D-D2D-D2H for Pinned Host Memory<-->Device Memory +/** + * Test Description + * ------------------------ + * - This performs the following scenarios of hipMemcpy2DAsync API on same GPU + 1. H2D-D2D-D2H for Host Memory<-->Device Memory + 2. H2D-D2D-D2H for Pinned Host Memory<-->Device Memory -Input : "A_h" initialized based on data type + Input : "A_h" initialized based on data type "A_h" --> "A_d" using H2D copy "A_d" --> "B_d" using D2D copy "B_d" --> "B_h" using D2H copy -Output: Validating A_h with B_h both should be equal for + Output: Validating A_h with B_h both should be equal for the number of COLUMNS and ROWS copied -*/ -TEMPLATE_TEST_CASE("Unit_hipMemcpy2DAsync_Host&PinnedMem", "", int, float, double) { - CHECK_IMAGE_SUPPORT + * Test source + * ------------------------ + * - unit/memory/hipMemcpy2DAsync.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 5.2 + */ +TEMPLATE_TEST_CASE("Unit_hipMemcpy2DAsync_Host&PinnedMem", "" + , int, float, double) { + CHECK_IMAGE_SUPPORT // 1 refers to pinned host memory auto mem_type = GENERATE(0, 1); HIP_CHECK(hipSetDevice(0)); @@ -108,9 +127,9 @@ TEMPLATE_TEST_CASE("Unit_hipMemcpy2DAsync_Host&PinnedMem", "", int, float, doubl hipMemcpyHostToDevice, hipStreamPerThread)); // Performs D2D on same GPU device - HIP_CHECK(hipMemcpy2DAsync(B_d, pitch_B, A_d, - pitch_A, COLUMNS*sizeof(TestType), - ROWS, hipMemcpyDeviceToDevice, hipStreamPerThread)); + HIP_CHECK(hipMemcpy2DAsync(B_d, pitch_B, A_d, pitch_A, + COLUMNS*sizeof(TestType), ROWS, + hipMemcpyDeviceToDevice, hipStreamPerThread)); // hipMemcpy2DAsync Device to Host HIP_CHECK(hipMemcpy2DAsync(B_h, COLUMNS*sizeof(TestType), B_d, pitch_B, @@ -136,22 +155,30 @@ TEMPLATE_TEST_CASE("Unit_hipMemcpy2DAsync_Host&PinnedMem", "", int, float, doubl HIP_CHECK(hipStreamDestroy(stream)); } -/* -This testcases performs the following scenarios of hipMemcpy2DAsync API -on Peer GPU -1. H2D-D2D-D2H for Host Memory<-->Device Memory -2. H2D-D2D-D2H for Pinned Host Memory<-->Device Memory +/** + * Test Description + * ------------------------ + * - This testcases performs the following scenarios of hipMemcpy2DAsync API on Peer GPU + 1. H2D-D2D-D2H for Host Memory<-->Device Memory + 2. H2D-D2D-D2H for Pinned Host Memory<-->Device Memory -Input : "A_h" initialized based on data type - "A_h" --> "A_d" using H2D copy - "A_d" --> "X_d" using D2D copy - "X_d" --> "B_h" using D2H copy -Output: Validating A_h with B_h both should be equal for - the number of COLUMNS and ROWS copied -*/ -TEMPLATE_TEST_CASE("Unit_hipMemcpy2DAsync_multiDevice-Host&PinnedMem", "", int, float, double) { + Input : "A_h" initialized based on data type + "A_h" --> "A_d" using H2D copy + "A_d" --> "X_d" using D2D copy + "X_d" --> "B_h" using D2H copy + Output: Validating A_h with B_h both should be equal for + the number of COLUMNS and ROWS copied + * Test source + * ------------------------ + * - unit/memory/hipMemcpy2DAsync.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 5.2 + */ + +TEMPLATE_TEST_CASE("Unit_hipMemcpy2DAsync_multiDevice-Host&PinnedMem", "" + , int, float, double) { CHECK_IMAGE_SUPPORT - auto mem_type = GENERATE(0, 1); int numDevices = 0; int canAccessPeer = 0; @@ -226,23 +253,31 @@ TEMPLATE_TEST_CASE("Unit_hipMemcpy2DAsync_multiDevice-Host&PinnedMem", "", int, } } -/* -This testcases performs the following scenarios of hipMemcpy2DAsync API -on Peer GPU -1. H2D-D2D-D2H for Host Memory<-->Device Memory -2. H2D-D2D-D2H for Pinned Host Memory<-->Device Memory -Memory is allocated in GPU-0 and Stream is created in GPU-1 +/** + * Test Description + * ------------------------ + * - This testcases performs the following scenarios of hipMemcpy2DAsync API on Peer GPU + 1. H2D-D2D-D2H for Host Memory<-->Device Memory + 2. H2D-D2D-D2H for Pinned Host Memory<-->Device Memory + Memory is allocated in GPU-0 and Stream is created in GPU-1 -Input : "A_h" initialized based on data type - "A_h" --> "A_d" using H2D copy - "A_d" --> "X_d" using D2D copy - "X_d" --> "B_h" using D2H copy -Output: Validating A_h with B_h both should be equal for - the number of COLUMNS and ROWS copied -*/ -TEMPLATE_TEST_CASE("Unit_hipMemcpy2DAsync_multiDevice-StreamOnDiffDevice", "", int, float, double) { + Input : "A_h" initialized based on data type + "A_h" --> "A_d" using H2D copy + "A_d" --> "X_d" using D2D copy + "X_d" --> "B_h" using D2H copy + Output: Validating A_h with B_h both should be equal for + the number of COLUMNS and ROWS copied + * Test source + * ------------------------ + * - unit/memory/hipMemcpy2DAsync.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 5.2 + */ + +TEMPLATE_TEST_CASE("Unit_hipMemcpy2DAsync_multiDevice-StreamOnDiffDevice", "" + , int, float, double) { CHECK_IMAGE_SUPPORT - auto mem_type = GENERATE(0, 1); int numDevices = 0; int canAccessPeer = 0; @@ -275,7 +310,6 @@ TEMPLATE_TEST_CASE("Unit_hipMemcpy2DAsync_multiDevice-StreamOnDiffDevice", "", i // Initialize the data HipTest::setDefaultData(NUM_W*NUM_H, A_h, B_h, C_h); - // Change device HIP_CHECK(hipSetDevice(1)); HIP_CHECK(hipStreamCreate(&stream)); @@ -284,7 +318,6 @@ TEMPLATE_TEST_CASE("Unit_hipMemcpy2DAsync_multiDevice-StreamOnDiffDevice", "", i HIP_CHECK(hipMemcpy2DAsync(A_d, pitch_A, A_h, COLUMNS*sizeof(TestType), COLUMNS*sizeof(TestType), ROWS, hipMemcpyHostToDevice, stream)); - // Device to Device HIP_CHECK(hipMemcpy2DAsync(X_d, pitch_X, A_d, pitch_A, COLUMNS*sizeof(TestType), @@ -318,12 +351,24 @@ TEMPLATE_TEST_CASE("Unit_hipMemcpy2DAsync_multiDevice-StreamOnDiffDevice", "", i } } -/* -This testcase verifies the null checks of hipMemcpy2DAsync API -*/ +/** + * Test Description + * ------------------------ + * - This testcase verifies the null checks of hipMemcpy2DAsync API + 1. hipMemcpy2DAsync API where Source Pitch is zero + 2. hipMemcpy2DAsync API where Destination Pitch is zero + 3. hipMemcpy2DAsync API where height is zero + 4. hipMemcpy2DAsync API where width is zero + * Test source + * ------------------------ + * - unit/memory/hipMemcpy2DAsync.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 5.2 + */ + TEST_CASE("Unit_hipMemcpy2DAsync_SizeCheck") { CHECK_IMAGE_SUPPORT - HIP_CHECK(hipSetDevice(0)); int* A_h{nullptr}, *A_d{nullptr}; size_t pitch_A; @@ -369,12 +414,23 @@ TEST_CASE("Unit_hipMemcpy2DAsync_SizeCheck") { free(A_h); } -/* -This testcase performs the negative scenarios of hipMemcpy2DAsync API -*/ +/** + * Test Description + * ------------------------ + * - This testcase performs the negative scenarios of hipMemcpy2DAsync API + 1. hipMemcpy2DAsync API by Passing nullptr to destination + 2. hipMemcpy2DAsync API by Passing nullptr to source + 3. hipMemcpy2DAsync API where width is > destination pitch + * Test source + * ------------------------ + * - unit/memory/hipMemcpy2DAsync.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 5.2 + */ + TEST_CASE("Unit_hipMemcpy2DAsync_Negative") { CHECK_IMAGE_SUPPORT - HIP_CHECK(hipSetDevice(0)); int* A_h{nullptr}, *A_d{nullptr}; size_t pitch_A; @@ -397,8 +453,8 @@ TEST_CASE("Unit_hipMemcpy2DAsync_Negative") { hipMemcpyDeviceToHost, stream) != hipSuccess); } - SECTION("hipMemcpy2DAsync API by Passing nullptr to destination") { - REQUIRE(hipMemcpy2DAsync(nullptr, width, nullptr, + SECTION("hipMemcpy2DAsync API by Passing nullptr to source") { + REQUIRE(hipMemcpy2DAsync(A_h, width, nullptr, pitch_A, COLUMNS*sizeof(int), ROWS, hipMemcpyDeviceToHost, stream) != hipSuccess); } @@ -414,3 +470,86 @@ TEST_CASE("Unit_hipMemcpy2DAsync_Negative") { HIP_CHECK(hipStreamDestroy(stream)); free(A_h); } + +static void hipMemcpy2DAsync_Basic_Size_Test(size_t inc) { + constexpr int defaultProgramSize = 256 * 1024 * 1024; + constexpr int N = 2; + constexpr int value = 42; + int *in, *out, *dev; + size_t newSize = 0, inp = 0; + size_t size = sizeof(int) * N * inc; + + size_t free, total; + HIP_CHECK(hipMemGetInfo(&free, &total)); + + if ( free < 2 * size ) + newSize = ( free - defaultProgramSize ) / 2; + else + newSize = size; + + INFO("Array size: " << size/1024.0/1024.0 << " MB or " << size << " Bytes."); + INFO("Free memory: " << free/1024.0/1024.0 << " MB or " << free << " Bytes"); + INFO("NewSize:" << newSize/1024.0/1024.0 << "MB or " << newSize << " Bytes"); + + HIP_CHECK(hipHostMalloc(&in, newSize)); + HIP_CHECK(hipHostMalloc(&out, newSize)); + HIP_CHECK(hipMalloc(&dev, newSize)); + + inp = newSize / (sizeof(int) * N); + for (size_t i=0; i < N; i++) { + in[i * inp] = value; + } + + size_t pitch = sizeof(int) * inp; + + hipStream_t stream; + HIP_CHECK(hipStreamCreate(&stream)); + + HIP_CHECK(hipMemcpy2DAsync(dev, pitch, in, pitch, sizeof(int), + N, hipMemcpyHostToDevice, stream)); + HIP_CHECK(hipMemcpy2DAsync(out, pitch, dev, pitch, sizeof(int), + N, hipMemcpyDeviceToHost, stream)); + HIP_CHECK(hipStreamSynchronize(stream)); + + for (size_t i=0; i < N; i++) { + REQUIRE(out[i * inp] == value); + } + + HIP_CHECK(hipFree(dev)); + HIP_CHECK(hipHostFree(in)); + HIP_CHECK(hipHostFree(out)); + HIP_CHECK(hipStreamDestroy(stream)); +} + +/** + * Test Description + * ------------------------ + * - This testcase performs multidevice size check on hipMemcpy2DAsync API + 1. Verify hipMemcpy2DAsync with 1 << 20 size + 2. Verify hipMemcpy2DAsync with 1 << 21 size + * Test source + * ------------------------ + * - unit/memory/hipMemcpy2DAsync.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + +TEST_CASE("Unit_hipMemcpy2DAsync_multiDevice_Basic_Size_Test") { + CHECK_IMAGE_SUPPORT + size_t input = 1 << 20; + int numDevices = 0; + HIP_CHECK(hipGetDeviceCount(&numDevices)); + + for (int i=0; i < numDevices; i++) { + HIP_CHECK(hipSetDevice(i)); + + SECTION("Verify hipMemcpy2DAsync with 1 << 20 size") { + hipMemcpy2DAsync_Basic_Size_Test(input); + } + SECTION("Verify hipMemcpy2DAsync with 1 << 21 size") { + input <<= 1; + hipMemcpy2DAsync_Basic_Size_Test(input); + } + } +} diff --git a/catch/unit/memory/hipMemcpy3DAsync_old.cc b/catch/unit/memory/hipMemcpy3DAsync_old.cc index 027df37ed9..264c31569b 100644 --- a/catch/unit/memory/hipMemcpy3DAsync_old.cc +++ b/catch/unit/memory/hipMemcpy3DAsync_old.cc @@ -1,5 +1,5 @@ /* -Copyright (c) 2021 Advanced Micro Devices, Inc. All rights reserved. +Copyright (c) 2021-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 @@ -17,6 +17,14 @@ OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE SOFTWARE. */ +/** + * @addtogroup hipMemcpy3DAsync hipMemcpy3DAsync + * @{ + * @ingroup MemoryTest + * `hipMemcpy3DAsync(const hipMemcpy3DParms* p, hipStream_t stream)` - + * Copies data between 3D objects. + */ + /* * This testfile verifies the following Scenarios of hipMemcpy3DAsync API @@ -653,13 +661,23 @@ void Memcpy3DAsync::simple_Memcpy3DAsync() { free(hOutputData); DeAllocateMemory(); } -/* -This testcase verifies hipMemcpyAsync for different datatypes -and different sizes -*/ -TEMPLATE_TEST_CASE("Unit_hipMemcpy3DAsync_Basic", "[hipMemcpy3DAsync]", int, unsigned int, float) { - CHECK_IMAGE_SUPPORT +/** + * Test Description + * ------------------------ + * - This testcase verifies hipMemcpyAsync for different datatypes and different sizes + * Test source + * ------------------------ + * - unit/memory/hipMemcpy3DAsync_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + +TEMPLATE_TEST_CASE("Unit_hipMemcpy3DAsync_Basic", + "[hipMemcpy3DAsync]", + int, unsigned int, float) { + CHECK_IMAGE_SUPPORT int numDevices = 0; HIP_CHECK(hipGetDeviceCount(&numDevices)); int device = -1; @@ -687,25 +705,39 @@ TEMPLATE_TEST_CASE("Unit_hipMemcpy3DAsync_Basic", "[hipMemcpy3DAsync]", int, uns } } -/* -This testcase performs the extent validation scenarios of -hipMemcpy3D API -*/ +/** + * Test Description + * ------------------------ + * - This testcase performs the extent validation scenarios of hipMemcpy3D API + * Test source + * ------------------------ + * - unit/memory/hipMemcpy3DAsync_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + TEST_CASE("Unit_hipMemcpy3DAsync_ExtentValidation") { CHECK_IMAGE_SUPPORT - Memcpy3DAsync memcpy3d(width, height, depth, hipChannelFormatKindSigned); memcpy3d.Extent_Validation(); } -/* -This testcase performs the negative scenarios of -hipMemcpy3DAsync API -*/ +/** + * Test Description + * ------------------------ + * - This testcase performs the negative scenarios of hipMemcpy3DAsync API + * Test source + * ------------------------ + * - unit/memory/hipMemcpy3DAsync_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + TEST_CASE("Unit_hipMemcpy3DAsync_multiDevice-Negative") { CHECK_IMAGE_SUPPORT - int numDevices = 0; HIP_CHECK(hipGetDeviceCount(&numDevices)); if (numDevices > 1) { @@ -717,13 +749,20 @@ TEST_CASE("Unit_hipMemcpy3DAsync_multiDevice-Negative") { } } -/* -This testcase performs the D2H,H2D and D2D on peer -GPU device -*/ +/** + * Test Description + * ------------------------ + * - This testcase performs the D2H,H2D and D2D on peer GPU device + * Test source + * ------------------------ + * - unit/memory/hipMemcpy3DAsync_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + TEST_CASE("Unit_hipMemcpy3DAsync_multiDevice-D2D") { CHECK_IMAGE_SUPPORT - int numDevices = 0; HIP_CHECK(hipGetDeviceCount(&numDevices)); if (numDevices > 1) { @@ -743,14 +782,21 @@ TEST_CASE("Unit_hipMemcpy3DAsync_multiDevice-D2D") { } } -/* -This testcase checks hipMemcpy3DAsync API by -allocating memory in one GPU and creating stream -in another GPU -*/ +/** + * Test Description + * ------------------------ + * - This testcase checks hipMemcpy3DAsync API by + allocating memory in one GPU and creating stream in another GPU + * Test source + * ------------------------ + * - unit/memory/hipMemcpy3DAsync_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + TEST_CASE("Unit_hipMemcpy3DAsync_multiDevice-DiffStream") { CHECK_IMAGE_SUPPORT - int numDevices = 0; HIP_CHECK(hipGetDeviceCount(&numDevices)); if (numDevices > 1) { @@ -761,3 +807,31 @@ TEST_CASE("Unit_hipMemcpy3DAsync_multiDevice-DiffStream") { SUCCEED("skipping the testcases as numDevices < 2"); } } + +/** + * Test Description + * ------------------------ + * - This testcase performs size check on hipMemcpy3DAsync API + 1. Verify with 128 for all height, width & depth value + * Test source + * ------------------------ + * - unit/memory/hipMemcpy3DAsync_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + +TEST_CASE("Unit_hipMemcpy3DAsync_Basic_Size_Test") { + CHECK_IMAGE_SUPPORT + constexpr int size_128b = 128; + int numDevices = 0; + HIP_CHECK(hipGetDeviceCount(&numDevices)); + + if (numDevices > 1) { + SECTION("Verify with 128 for all height, width & depth value") { + Memcpy3DAsync memcpy3dAsync(size_128b, size_128b, size_128b, + hipChannelFormatKindFloat); + memcpy3dAsync.D2D_SameDeviceMem_StreamDiffDevice(); + } + } +} diff --git a/catch/unit/memory/hipMemcpy3D_old.cc b/catch/unit/memory/hipMemcpy3D_old.cc index 5a75bb5a70..1fb32183a8 100644 --- a/catch/unit/memory/hipMemcpy3D_old.cc +++ b/catch/unit/memory/hipMemcpy3D_old.cc @@ -1,5 +1,5 @@ /* -Copyright (c) 2021 Advanced Micro Devices, Inc. All rights reserved. +Copyright (c) 2021-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 @@ -17,6 +17,14 @@ OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN THE SOFTWARE. */ +/** + * @addtogroup hipMemcpy3D hipMemcpy3D + * @{ + * @ingroup MemoryTest + * `hipMemcpy3D(const hipMemcpy3DParms* p)` - + * Copies data between 3D objects. + */ + /* * This testfile verifies the following scenarios of hipMemcpy3D API * @@ -38,14 +46,14 @@ static constexpr auto depth{10}; template class Memcpy3D { - int width, height, depth; - unsigned int size; + size_t width, height, depth; + size_t size; hipArray_t arr, arr1; hipChannelFormatKind formatKind; hipMemcpy3DParms myparms; T* hData; public: - Memcpy3D(int l_width, int l_height, int l_depth, + Memcpy3D(size_t l_width, size_t l_height, size_t l_depth, hipChannelFormatKind l_format); void simple_Memcpy3D(); void Extent_Validation(); @@ -71,7 +79,7 @@ void Memcpy3D::SetDefaultData() { * Constructor initalized width,depth and height */ template -Memcpy3D::Memcpy3D(int l_width, int l_height, int l_depth, +Memcpy3D::Memcpy3D(size_t l_width, size_t l_height, size_t l_depth, hipChannelFormatKind l_format) { width = l_width; height = l_height; @@ -546,17 +554,27 @@ void Memcpy3D::simple_Memcpy3D() { free(hOutputData); DeAllocateMemory(); } -/* - This testcase performs hipMemcpy3D API validation for - different datatypes and different sizes -*/ -TEMPLATE_TEST_CASE("Unit_hipMemcpy3D_Basic", "[hipMemcpy3D]", int, unsigned int, float) { - CHECK_IMAGE_SUPPORT +/** + * Test Description + * ------------------------ + * - This testcase performs hipMemcpy3D API validation for + different datatypes and different sizes + * Test source + * ------------------------ + * - unit/memory/hipMemcpy3D_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 5.2 + */ + +TEMPLATE_TEST_CASE("Unit_hipMemcpy3D_Basic", "[hipMemcpy3D]", + int, unsigned int, float) { + CHECK_IMAGE_SUPPORT int device = -1; HIP_CHECK(hipGetDevice(&device)); hipDeviceProp_t prop; - HIP_CHECK(hipGetDeviceProperties(&prop,device)); + HIP_CHECK(hipGetDeviceProperties(&prop, device)); auto i = GENERATE_COPY(10, 100, 1024, prop.maxTexture3D[0]); auto j = GENERATE(10, 100); int numDevices = 0; @@ -577,25 +595,39 @@ TEMPLATE_TEST_CASE("Unit_hipMemcpy3D_Basic", "[hipMemcpy3D]", int, unsigned int, } } -/* -This testcase performs the extent validation scenarios of -hipMemcpy3D API -*/ +/** + * Test Description + * ------------------------ + * - This testcase performs the extent validation scenarios of hipMemcpy3D API + * Test source + * ------------------------ + * - unit/memory/hipMemcpy3D_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 5.2 + */ + TEST_CASE("Unit_hipMemcpy3D_ExtentValidation") { CHECK_IMAGE_SUPPORT - Memcpy3D memcpy3d(width, height, depth, hipChannelFormatKindSigned); memcpy3d.Extent_Validation(); } -/* -This testcase performs the negative scenarios of -hipMemcpy3D API -*/ +/** + * Test Description + * ------------------------ + * - This testcase performs the negative scenarios of hipMemcpy3D API + * Test source + * ------------------------ + * - unit/memory/hipMemcpy3D_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 5.2 + */ + TEST_CASE("Unit_hipMemcpy3D_multiDevice-Negative") { CHECK_IMAGE_SUPPORT - int numDevices = 0; HIP_CHECK(hipGetDeviceCount(&numDevices)); if (numDevices > 1) { @@ -607,13 +639,22 @@ TEST_CASE("Unit_hipMemcpy3D_multiDevice-Negative") { } } -/* -This testcase performs the D2H,H2D and D2D on peer -GPU device -*/ +/** + * Test Description + * ------------------------ + * - This testcase performs the D2H,H2D and D2D on peer GPU device + 1. Verify with D2H & H2D On DiffDevice + 2. Verify with D2D On DiffDevice + * Test source + * ------------------------ + * - unit/memory/hipMemcpy3D_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 5.2 + */ + TEST_CASE("Unit_hipMemcpy3D_multiDevice-OnPeerDevice") { CHECK_IMAGE_SUPPORT - int numDevices = 0; HIP_CHECK(hipGetDeviceCount(&numDevices)); if (numDevices > 1) { @@ -632,3 +673,51 @@ TEST_CASE("Unit_hipMemcpy3D_multiDevice-OnPeerDevice") { SUCCEED("skipping the testcases as numDevices < 2"); } } + +/** + * Test Description + * ------------------------ + * - This testcase performs multidevice size check on hipMemcpy3D API + 1. Verify with 128 for all height, width & depth value + 2. Verify with 256 for height and 128 for width & depth value + 3. Verify with 256 for width and 128 for height & depth value + 4. Verify with 256 for depth and 128 for height & width value + * Test source + * ------------------------ + * - unit/memory/hipMemcpy3D_old.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ + +TEST_CASE("Unit_hipMemcpy3D_multiDevice_Basic_Size_Test") { + CHECK_IMAGE_SUPPORT + constexpr int size_128b = 128, size_256b = 256; + int numDevices = 0; + HIP_CHECK(hipGetDeviceCount(&numDevices)); + + for (int i=0; i < numDevices; i++) { + HIP_CHECK(hipSetDevice(i)); + + SECTION("Verify with 128 for all height, width & depth value") { + Memcpy3D memcpy3d_obj1(size_128b, size_128b, size_128b, + hipChannelFormatKindUnsigned); + memcpy3d_obj1.simple_Memcpy3D(); + } + SECTION("Verify with 256 for height and 128 for width & depth value") { + Memcpy3D memcpy3d_obj2(size_256b, size_128b, size_128b, + hipChannelFormatKindUnsigned); + memcpy3d_obj2.simple_Memcpy3D(); + } + SECTION("Verify with 256 for width and 128 for height & depth value") { + Memcpy3D memcpy3d_obj3(size_128b, size_256b, size_128b, + hipChannelFormatKindFloat); + memcpy3d_obj3.simple_Memcpy3D(); + } + SECTION("Verify with 256 for depth and 128 for height & width value") { + Memcpy3D memcpy3d_obj4(size_128b, size_128b, size_256b, + hipChannelFormatKindUnsigned); + memcpy3d_obj4.simple_Memcpy3D(); + } + } +}