From f4697904aeb6e12455ec89aeba82d4511ef41171 Mon Sep 17 00:00:00 2001 From: "Yendluri, Manas" Date: Thu, 14 Mar 2024 17:22:31 +0530 Subject: [PATCH] SWDEV-448562 - [catch2][dtest] Funtional and negative testcases for hipDrvMemcpy2DUnaligned API Change-Id: I0761b9ecac39bf429cd7943a13bb77d8784b3524 --- catch/unit/memory/CMakeLists.txt | 2 + catch/unit/memory/hipDrvMemcpy2DUnaligned.cc | 254 +++++++++++++++++++ 2 files changed, 256 insertions(+) create mode 100644 catch/unit/memory/hipDrvMemcpy2DUnaligned.cc diff --git a/catch/unit/memory/CMakeLists.txt b/catch/unit/memory/CMakeLists.txt index e934cbee5f..cdd94b0e92 100644 --- a/catch/unit/memory/CMakeLists.txt +++ b/catch/unit/memory/CMakeLists.txt @@ -168,6 +168,8 @@ set(TEST_SRC if(HIP_PLATFORM MATCHES "amd") set(TEST_SRC ${TEST_SRC} + # Test below currently doesn't work on NVIDIA, will update based on ticket response + hipDrvMemcpy2DUnaligned.cc # Below 3 tests should be compiled for NVIDIA as well after EXSWHTEC-349 is addressed hipArrayGetInfo.cc hipArrayGetDescriptor.cc diff --git a/catch/unit/memory/hipDrvMemcpy2DUnaligned.cc b/catch/unit/memory/hipDrvMemcpy2DUnaligned.cc new file mode 100644 index 0000000000..b5d2fb70f3 --- /dev/null +++ b/catch/unit/memory/hipDrvMemcpy2DUnaligned.cc @@ -0,0 +1,254 @@ +/* +Copyright (c) 2024 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 +/** + * @addtogroup hipDrvMemcpy2DUnaligned hipDrvMemcpy2DUnaligned + * @{ + * @ingroup MemoryTest + * `hipError_t hipDrvMemcpy2DUnaligned(const hip_Memcpy2D * pCopy)` - + * Copies memory for 2D arrays. + */ + +/** + * Test Description + * ------------------------ + * - Negative test cases for hipDrvMemcpy2DUnaligned api + * Test source + * ------------------------ + * - unit/memory/hipDrvMemcpy2DUnaligned.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ +TEST_CASE("Unit_hipDrvMemcpy2DUnaligned_NegTst") { + // declare host and device arrays + int rows, cols; + rows = GENERATE(3, 4, 100); + cols = GENERATE(3, 4, 100); + int *srcD, *srcH; + int *dstD, *dstH; + HIP_CHECK(hipMalloc(reinterpret_cast(&srcD), + sizeof(int) * rows * cols)); + HIP_CHECK(hipMalloc(reinterpret_cast(&dstD), + sizeof(int) * rows * cols)); + HIP_CHECK(hipHostMalloc(reinterpret_cast(&srcH), + sizeof(int) * rows * cols)); + HIP_CHECK(hipHostMalloc(reinterpret_cast(&dstH), + sizeof(int) * rows * cols)); + + // initialise array with corresponding index values + for (int i = 0; i < rows * cols; i++) { + srcH[i] = i; + } + HIP_CHECK(hipMemcpyHtoD(srcD, srcH, sizeof(int) * rows * cols)); + + hip_Memcpy2D pCopy; + + SECTION("srcY(second argument) + non-zero WidthInBytes(15th argument)\ + * Height(16th argument) points to unallocated memory") { + pCopy.srcXInBytes = 0; + pCopy.srcY = rows; + pCopy.srcMemoryType = hipMemoryTypeDevice; + pCopy.dstXInBytes = 0; + pCopy.dstY = 0; + pCopy.dstMemoryType = hipMemoryTypeDevice; + pCopy.srcDevice = srcD; + pCopy.dstDevice = dstD; + pCopy.WidthInBytes = sizeof(int); + pCopy.Height = 1; + pCopy.srcPitch = cols * sizeof(int); + pCopy.dstPitch = cols * sizeof(int); + HIP_CHECK_ERROR(hipDrvMemcpy2DUnaligned(&pCopy), hipErrorInvalidValue); + } + SECTION("srcHost(4th argument), srcDevice(5th argument), srcArray(6th\ + argument) passed nullptr") { + pCopy.srcXInBytes = 0; + pCopy.srcY = 0; + pCopy.dstXInBytes = 0; + pCopy.dstY = 0; + pCopy.dstMemoryType = hipMemoryTypeDevice; + pCopy.dstDevice = dstD; + pCopy.WidthInBytes = sizeof(int) * cols; + pCopy.Height = rows; + pCopy.srcPitch = cols * sizeof(int); + pCopy.dstPitch = cols * sizeof(int); + SECTION("srcHost passed nullptr") { + pCopy.srcMemoryType = hipMemoryTypeHost; + pCopy.srcHost = nullptr; + } + SECTION("srcDevice passed nullptr") { + pCopy.srcMemoryType = hipMemoryTypeDevice; + pCopy.srcDevice = nullptr; + } + SECTION("srcArray passed nullptr") { + pCopy.srcMemoryType = hipMemoryTypeArray; + pCopy.srcArray = nullptr; + } + HIP_CHECK_ERROR(hipDrvMemcpy2DUnaligned(&pCopy), hipErrorInvalidValue); + } + SECTION("dstY(second argument) + non-zero WidthInBytes(15th argument)\ + * Height(16th argument) points to unallocated memory") { + pCopy.srcXInBytes = 0; + pCopy.srcY = 0; + pCopy.srcMemoryType = hipMemoryTypeDevice; + pCopy.dstXInBytes = 0; + pCopy.dstY = rows; + pCopy.dstMemoryType = hipMemoryTypeDevice; + pCopy.srcDevice = srcD; + pCopy.dstDevice = dstD; + pCopy.WidthInBytes = sizeof(int); + pCopy.Height = 1; + pCopy.srcPitch = cols * sizeof(int); + pCopy.dstPitch = cols * sizeof(int); + HIP_CHECK_ERROR(hipDrvMemcpy2DUnaligned(&pCopy), hipErrorInvalidValue); + } + SECTION("dstHost(4th argument), dstDevice(5th argument), dstArray(6th\ + argument) passed nullptr") { + pCopy.srcXInBytes = 0; + pCopy.srcY = 0; + pCopy.srcMemoryType = hipMemoryTypeDevice; + pCopy.srcDevice = srcD; + pCopy.dstXInBytes = 0; + pCopy.dstY = 0; + pCopy.WidthInBytes = sizeof(int) * cols; + pCopy.Height = rows; + pCopy.srcPitch = cols * sizeof(int); + pCopy.dstPitch = cols * sizeof(int); + SECTION("dstHost passed nullptr") { + pCopy.dstMemoryType = hipMemoryTypeHost; + pCopy.dstHost = nullptr; + } + SECTION("dstDevice passed nullptr") { + pCopy.dstMemoryType = hipMemoryTypeDevice; + pCopy.dstDevice = nullptr; + } + SECTION("dstArray passed nullptr") { + pCopy.dstMemoryType = hipMemoryTypeArray; + pCopy.dstArray = nullptr; + } + HIP_CHECK_ERROR(hipDrvMemcpy2DUnaligned(&pCopy), hipErrorInvalidValue); + } + SECTION("WidthInBytes * Height greater than allocated memory(both src \ + and dst)") { + pCopy.srcXInBytes = 0; + pCopy.srcY = 0; + pCopy.srcMemoryType = hipMemoryTypeDevice; + pCopy.dstXInBytes = 0; + pCopy.dstY = 0; + pCopy.dstMemoryType = hipMemoryTypeDevice; + pCopy.srcDevice = srcD; + pCopy.dstDevice = dstD; + pCopy.WidthInBytes = sizeof(int) * cols; + pCopy.Height = rows + 1; + pCopy.srcPitch = cols * sizeof(int); + pCopy.dstPitch = cols * sizeof(int); + HIP_CHECK_ERROR(hipDrvMemcpy2DUnaligned(&pCopy), hipErrorInvalidValue); + } + HIP_CHECK(hipFree(srcH)); + HIP_CHECK(hipFree(srcD)); + HIP_CHECK(hipFree(dstD)); + HIP_CHECK(hipFree(dstH)); +} + +/** + * Test Description + * ------------------------ + * - Functional test cases for hipDrvMemcpy2DUnaligned api + * Test source + * ------------------------ + * - unit/memory/hipDrvMemcpy2DUnaligned.cc + * Test requirements + * ------------------------ + * - HIP_VERSION >= 6.0 + */ +TEST_CASE("Unit_hipDrvMemcpy2DUnaligned_FuncTst") { + SECTION("Different types of memory transfers functional tests to check if\ + copied array contains correct values") { + // declare host and device arrays + int rows, cols; + rows = GENERATE(3, 4, 100); + cols = GENERATE(3, 4, 100); + int *srcD, *srcH; + int *dstD, *dstH; + HIP_CHECK(hipMalloc(reinterpret_cast(&srcD), + sizeof(int) * rows * cols)); + HIP_CHECK(hipMalloc(reinterpret_cast(&dstD), + sizeof(int) * rows * cols)); + srcH = reinterpret_cast(malloc(sizeof(int) * rows * cols)); + dstH = reinterpret_cast(malloc(sizeof(int) * rows * cols)); + + // initialise array with corresponding index values + for (int i = 0; i < rows * cols; i++) { + srcH[i] = i; + } + + hip_Memcpy2D pCopy; + pCopy.srcXInBytes = 0; + pCopy.srcY = 0; + pCopy.dstXInBytes = 0; + pCopy.dstY = 0; + pCopy.WidthInBytes = sizeof(int) * cols; + pCopy.Height = rows; + pCopy.srcPitch = cols * sizeof(int); + pCopy.dstPitch = cols * sizeof(int); + + SECTION("Device to Device") { + HIP_CHECK(hipMemcpyHtoD(srcD, srcH, sizeof(int) * rows * cols)); + pCopy.srcMemoryType = hipMemoryTypeDevice; + pCopy.dstMemoryType = hipMemoryTypeDevice; + pCopy.srcDevice = srcD; + pCopy.dstDevice = dstD; + HIP_CHECK(hipDrvMemcpy2DUnaligned(&pCopy)); + HIP_CHECK(hipMemcpyDtoH(dstH, dstD, sizeof(int) * rows * cols)); + } + SECTION("Device to Host") { + HIP_CHECK(hipMemcpyHtoD(srcD, srcH, sizeof(int) * rows * cols)); + pCopy.srcMemoryType = hipMemoryTypeDevice; + pCopy.dstMemoryType = hipMemoryTypeHost; + pCopy.srcDevice = srcD; + pCopy.dstHost = dstH; + HIP_CHECK(hipDrvMemcpy2DUnaligned(&pCopy)); + } + SECTION("Host to Device") { + pCopy.srcMemoryType = hipMemoryTypeHost; + pCopy.dstMemoryType = hipMemoryTypeDevice; + pCopy.srcHost = srcH; + pCopy.dstDevice = dstD; + HIP_CHECK(hipDrvMemcpy2DUnaligned(&pCopy)); + HIP_CHECK(hipMemcpyDtoH(dstH, dstD, sizeof(int) * rows * cols)); + } + SECTION("Host to Host") { + pCopy.srcMemoryType = hipMemoryTypeHost; + pCopy.dstMemoryType = hipMemoryTypeHost; + pCopy.srcHost = srcH; + pCopy.dstHost = dstH; + HIP_CHECK(hipDrvMemcpy2DUnaligned(&pCopy)); + } + + for (int i = 0; i < rows * cols; i++) { + REQUIRE(dstH[i] == i); + } + + HIP_CHECK(hipFree(srcD)); + HIP_CHECK(hipFree(dstD)); + free(srcH); + free(dstH); + } +}