SWDEV-416449 - [catch2][dtest] Functional test cases for the memory type attribute of hipPointerGetAttributes() API

Change-Id: I7d0dc9f9a646ee3ea5aaa0d7f416817594ad7409

catch/unit/memory/hipPointerGetAttributes.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
@@ -28,27 +28,40 @@ Following scenarios are verified for hipPointerGetAttributes API
 4. Multi-threaded test with many simul allocs.
 
 */
+#include <inttypes.h>
 #include <hip_test_common.hh>
+#include <hip_test_defgroups.hh>
+#ifdef __linux__
+#include <sys/mman.h>
+#endif
 #include <vector>
 #include <iostream>
 #include <string>
-
 size_t Nbytes = 0;
 constexpr size_t N{1000000};
-
+/**
+ * @addtogroup hipPointerGetAttributes hipPointerGetAttributes
+ * @{
+ * @ingroup MemoryTest
+ * `hipError_t  hipPointerGetAttributes (hipPointerAttribute_t *attributes, const void *ptr)` -
+ * Queries the memory pointer attributes.
+ */
 
 //=================================================================================================
 // Utility Functions:
 //=================================================================================================
 
-bool operator==(const hipPointerAttribute_t& lhs, const hipPointerAttribute_t& rhs) {
-  return ((lhs.hostPointer == rhs.hostPointer) && (lhs.devicePointer == rhs.devicePointer) &&
+bool operator==(const hipPointerAttribute_t& lhs,
+                const hipPointerAttribute_t& rhs) {
+  return ((lhs.hostPointer == rhs.hostPointer) &&
+          (lhs.devicePointer == rhs.devicePointer) &&
           (lhs.type == rhs.type) && (lhs.device == rhs.device) &&
           (lhs.allocationFlags == rhs.allocationFlags));
 }
 
 
-bool operator!=(const hipPointerAttribute_t& lhs, const hipPointerAttribute_t& rhs) {
+bool operator!=(const hipPointerAttribute_t& lhs,
+                const hipPointerAttribute_t& rhs) {
   return !(lhs == rhs);
 }
 
@@ -79,7 +92,8 @@ void printAttribs(const hipPointerAttribute_t* attribs) {
   printf(
       "hostPointer:%p devicePointer:%p  type:%s deviceId:%d isManaged:%d "
       "allocationFlags:%u\n",
-      attribs->hostPointer, attribs->devicePointer, memoryTypeToString(attribs->type),
+      attribs->hostPointer, attribs->devicePointer,
+      memoryTypeToString(attribs->type),
       attribs->device, attribs->isManaged, attribs->allocationFlags);
 }
 
@@ -98,7 +112,8 @@ struct SuperPointerAttribute {
 
 
 // Support function to check result against a reference:
-void checkPointer(const SuperPointerAttribute& ref, int major, int minor, void* pointer) {
+void checkPointer(const SuperPointerAttribute& ref, int major,
+                  int minor, void* pointer) {
   hipPointerAttribute_t attribs;
   resetAttribs(&attribs);
 
@@ -122,8 +137,9 @@ void checkPointer(const SuperPointerAttribute& ref, int major, int minor, void*
 // we do this in the testMultiThreaded_1 test.
 void clusterAllocs(int numAllocs, size_t minSize, size_t maxSize) {
   Nbytes = N * sizeof(char);
-  printf("clusterAllocs numAllocs=%d size=%lu..%lu\n",
-         numAllocs, static_cast<unsigned long>(minSize), static_cast<unsigned long>(maxSize));
+  printf("clusterAllocs numAllocs=%d size=" "%" PRIu64 ".." "%" PRIu64 "\n",
+         numAllocs, static_cast<uint64_t>(minSize),
+         static_cast<uint64_t>(maxSize));
   const int Max_Devices = 256;
   std::vector<SuperPointerAttribute> reference(numAllocs);
 
@@ -150,14 +166,17 @@ void clusterAllocs(int numAllocs, size_t minSize, size_t maxSize) {
 
     void* ptr;
     if (isDevice) {
-      totalDeviceAllocated[reference[i]._attrib.device] += reference[i]._sizeBytes;
-      HIP_CHECK(hipMalloc(reinterpret_cast<void**>(&ptr), reference[i]._sizeBytes));
+      totalDeviceAllocated[reference[i]._attrib.device] +=
+                                  reference[i]._sizeBytes;
+      HIP_CHECK(hipMalloc(reinterpret_cast<void**>(&ptr),
+                               reference[i]._sizeBytes));
       reference[i]._attrib.type = hipMemoryTypeDevice;
       reference[i]._attrib.devicePointer = ptr;
       reference[i]._attrib.hostPointer = NULL;
       reference[i]._attrib.allocationFlags = 0;
     } else {
-      HIP_CHECK(hipHostMalloc(reinterpret_cast<void**>(&ptr), reference[i]._sizeBytes,
+      HIP_CHECK(hipHostMalloc(reinterpret_cast<void**>(&ptr),
+                              reference[i]._sizeBytes,
                               hipHostMallocDefault));
       reference[i]._attrib.type = hipMemoryTypeHost;
       reference[i]._attrib.devicePointer = ptr;
@@ -173,10 +192,12 @@ void clusterAllocs(int numAllocs, size_t minSize, size_t maxSize) {
     HIP_CHECK(hipMemGetInfo(&free, &total));
     printf(
         "  device#%d: hipMemGetInfo: "
-        "free=%zu (%4.2fMB) totalDevice=%lu (%4.2fMB) total=%zu "
+        "free=%zu (%4.2fMB) totalDevice=""%" PRIu64 "(%4.2fMB) total=%zu "
         "(%4.2fMB)\n",
-        i, free, (free / 1024.0 / 1024.0), static_cast<unsigned long>(totalDeviceAllocated[i]),
-        (totalDeviceAllocated[i]) / 1024.0 / 1024.0, total, (total / 1024.0 / 1024.0));
+        i, free, (free / 1024.0 / 1024.0),
+        static_cast<uint64_t>(totalDeviceAllocated[i]),
+        (totalDeviceAllocated[i]) / 1024.0 / 1024.0, total,
+        (total / 1024.0 / 1024.0));
     REQUIRE(free + totalDeviceAllocated[i] <= total);
   }
 
@@ -198,12 +219,17 @@ void clusterAllocs(int numAllocs, size_t minSize, size_t maxSize) {
     }
   }
 }
-
-//========================================================================
-// Functions to run tests
-//=======================================================================
-//--
-// Run through a couple simple cases to test lookups host pointer arithmetic:
+/**
+ * Test Description
+ * ------------------------
+ * - Run through a couple simple cases to test lookups host pointer arithmetic
+// specified size range
+ * ------------------------
+ *  - unit/memory/hipPointerGetAttributes.cc
+ * Test requirements
+ * ------------------------
+ *  - HIP_VERSION >= 5.7
+ */
 TEST_CASE("Unit_hipPointerGetAttributes_Basic") {
   HIP_CHECK(hipSetDevice(0));
   Nbytes = N * sizeof(char);
@@ -217,12 +243,15 @@ TEST_CASE("Unit_hipPointerGetAttributes_Basic") {
   hipError_t e;
 
   HIP_CHECK(hipMalloc(&A_d, Nbytes));
-  HIP_CHECK(hipHostMalloc(reinterpret_cast<void**>(&A_Pinned_h), Nbytes, hipHostMallocDefault));
+  HIP_CHECK(hipHostMalloc(reinterpret_cast<void**>(&A_Pinned_h),
+                          Nbytes, hipHostMallocDefault));
 
   size_t free, total;
   HIP_CHECK(hipMemGetInfo(&free, &total));
-  printf("hipMemGetInfo: free=%zu (%4.2f) Nbytes=%lu total=%zu (%4.2f)\n", free,
-         (free / 1024.0 / 1024.0), static_cast<unsigned long>(Nbytes), total, (total / 1024.0 / 1024.0));
+  printf("hipMemGetInfo: free=%zu (%4.2f) Nbytes=""%" PRIu64
+         "total=%zu (%4.2f)\n", free,
+         (free / 1024.0 / 1024.0), static_cast<uint64_t>(Nbytes),
+          total, (total / 1024.0 / 1024.0));
   REQUIRE(free + Nbytes <= total);
 
 
@@ -242,14 +271,16 @@ TEST_CASE("Unit_hipPointerGetAttributes_Basic") {
   // Corner case at end of array:
   resetAttribs(&attribs2);
   HIP_CHECK(hipPointerGetAttributes(&attribs2, A_d + Nbytes - 1));
-  REQUIRE((ptr + Nbytes - 1) == reinterpret_cast<char*>(attribs2.devicePointer));
+  REQUIRE((ptr + Nbytes - 1) == reinterpret_cast<char*>
+                               (attribs2.devicePointer));
 
   // Pointer just beyond array must be invalid or at least a different pointer
   resetAttribs(&attribs2);
   e = hipPointerGetAttributes(&attribs2, A_d + Nbytes + 1);
   if (e != hipErrorInvalidValue) {
     // We might have strayed into another pointer area.
-    REQUIRE(reinterpret_cast<char*>(ptr) != reinterpret_cast<char*>(attribs2.devicePointer));
+    REQUIRE(reinterpret_cast<char*>(ptr) !=
+            reinterpret_cast<char*>(attribs2.devicePointer));
   }
 
 
@@ -274,12 +305,34 @@ TEST_CASE("Unit_hipPointerGetAttributes_Basic") {
   // OS memory
   printf("\nOS-allocated memory (malloc)\n");
 }
+/**
+ * Test Description
+ * ------------------------
+ * - Test that allocates memory across all devices withing the specified size range
+ * Test source
+ * ------------------------
+ *  - unit/memory/hipPointerGetAttributes.cc
+ * Test requirements
+ * ------------------------
+ *  - HIP_VERSION >= 5.7
+ */
 
 TEST_CASE("Unit_hipPointerGetAttributes_ClusterAlloc") {
   srand(0x100);
   printf("\n=============================================\n");
   clusterAllocs(100, 1024 * 1, 1024 * 1024);
 }
+/**
+ * Test Description
+ * ------------------------
+ * - Test that allocates memory across all devices withing the specified size range
+ * Test source
+ * ------------------------
+ *  - unit/memory/hipPointerGetAttributes.cc
+ * Test requirements
+ * ------------------------
+ *  - HIP_VERSION >= 5.7
+ */
 
 TEST_CASE("Unit_hipPointerGetAttributes_TinyClusterAlloc") {
   srand(0x200);
@@ -289,7 +342,8 @@ TEST_CASE("Unit_hipPointerGetAttributes_TinyClusterAlloc") {
 
 // Multi-threaded test with many simul allocs.
 // IN : serialize will force the test to run in serial fashion.
-#if 0  // FIXME_jatinx These need to be ported to HIP_CHECK_THREAD. Disabling it for now
+#if 0  // FIXME_jatinx These need to be ported to HIP_CHECK_THREAD.
+Disabling it for now
 TEST_CASE("Unit_hipPointerGetAttributes_MultiThread") {
   srand(0x300);
   auto serialize = 1;
@@ -310,55 +364,77 @@ TEST_CASE("Unit_hipPointerGetAttributes_MultiThread") {
   if (serialize) t4.join();
 }
 #endif
+/**
+ * Test Description
+ * ------------------------
+ *  - Perform negative tests for the API hipPointerGetAttributes
+ * by passing invalid memory pointer and invalid attributes
+ * Test source
+ * ------------------------
+ *  - unit/memory/hipPointerGetAttributes.cc
+ * Test requirements
+ * ------------------------
+ *  - HIP_VERSION >= 5.7
+ */
 
 TEST_CASE("Unit_hipPointerGetAttributes_Negative") {
 #if HT_AMD  // Nvidia crashed in hipPointerGetAttributes on nullptr
   SECTION("Invalid Attributes Pointer") {
     int* dPtr{nullptr};
     HIP_CHECK(hipMalloc(&dPtr, sizeof(int)));
-    HIP_CHECK_ERROR(hipPointerGetAttributes(nullptr, dPtr), hipErrorInvalidValue);
+    HIP_CHECK_ERROR(hipPointerGetAttributes(nullptr, dPtr),
+                                     hipErrorInvalidValue);
     HIP_CHECK(hipFree(dPtr));
   }
 #endif
 
   SECTION("Invalid Device Pointer") {
     hipPointerAttribute_t attributes{};
-    HIP_CHECK_ERROR(hipPointerGetAttributes(&attributes, nullptr), hipErrorInvalidValue);
+    HIP_CHECK_ERROR(hipPointerGetAttributes(&attributes, nullptr),
+                                            hipErrorInvalidValue);
   }
 }
 
-// Run this test for all devices for DeviceMemory, HostMemory, ManagedMemory and MappedMemory
+/**
+ * Test Description
+ * ------------------------
+ *  - Run this test for all devices for DeviceMemory,
+ *    HostMemory and MappedMemory
+ * Test source
+ * ------------------------
+ *  - unit/memory/hipPointerGetAttributes.cc
+ * Test requirements
+ * ------------------------
+ *  - HIP_VERSION >= 6.0
+ */
 TEST_CASE("Unit_hipPointerGetAttributes_GpuIter") {
   int deviceCount{0};
   HIP_CHECK(hipGetDeviceCount(&deviceCount));
   REQUIRE(deviceCount != 0);
 
   // Memory Types
-  enum MemoryTypes { DeviceMemory = 0, HostMemory, MappedMemory };
+  enum MemoryTypes {DeviceMemory = 0, HostMemory, MappedMemory};
   auto MemoryType =
-      GENERATE(MemoryTypes::DeviceMemory, MemoryTypes::HostMemory, MemoryTypes::MappedMemory);
+      GENERATE(MemoryTypes::DeviceMemory, MemoryTypes::HostMemory,
+               MemoryTypes::MappedMemory);
 
   INFO("Memory Type: " << MemoryType);
 
   int* ptr{nullptr};
   for (int i = 0; i < deviceCount; i++) {
     HIP_CHECK(hipSetDevice(i));
-
     ptr = nullptr;
     if (MemoryType == MemoryTypes::DeviceMemory) {
       HIP_CHECK(hipMalloc(&ptr, sizeof(int)));
     } else if (MemoryType == MemoryTypes::HostMemory) {
-      HIP_CHECK(hipHostMalloc(&ptr, sizeof(int)));
+        HIP_CHECK(hipHostMalloc(&ptr, sizeof(int)));
     } else if (MemoryType == MemoryTypes::MappedMemory) {
-      HIP_CHECK(hipHostMalloc(&ptr, sizeof(int), hipHostMallocMapped));
+        HIP_CHECK(hipHostMalloc(&ptr, sizeof(int), hipHostMallocMapped));
     }
-
     REQUIRE(ptr != nullptr);
     hipPointerAttribute_t attributes{};
     HIP_CHECK(hipPointerGetAttributes(&attributes, ptr));
-
     REQUIRE(attributes.device == i);  // Device Check
-
     // Memory address and type check
     if (MemoryType == MemoryTypes::DeviceMemory) {
       REQUIRE(attributes.type == hipMemoryTypeDevice);
@@ -369,15 +445,116 @@ TEST_CASE("Unit_hipPointerGetAttributes_GpuIter") {
       REQUIRE(attributes.hostPointer == ptr);
     } else if (MemoryType == MemoryTypes::MappedMemory) {
       int* devicePtr{nullptr};
-      HIP_CHECK(hipHostGetDevicePointer(reinterpret_cast<void**>(&devicePtr), ptr, 0));
+      HIP_CHECK(hipHostGetDevicePointer(reinterpret_cast<void**>(&devicePtr),
+                                        ptr, 0));
       REQUIRE(attributes.hostPointer == ptr);
       REQUIRE(attributes.hostPointer == devicePtr);
     }
 
     if (MemoryType == MemoryTypes::DeviceMemory) {
       HIP_CHECK(hipFree(ptr));
-    } else if (MemoryType == MemoryTypes::MappedMemory || MemoryType == MemoryTypes::HostMemory) {
+    } else if (MemoryType == MemoryTypes::MappedMemory ||
+               MemoryType == MemoryTypes::HostMemory) {
       HIP_CHECK(hipHostFree(ptr));
     }
   }
 }
+/**
+ * Test Description
+ * ------------------------
+ *  - Managed memory type attribute verification.
+ * Test source
+ * ------------------------
+ *  - unit/memory/hipPointerGetAttributes.cc
+ * Test requirements
+ * ------------------------
+ *  - HIP_VERSION >= 6.0
+ */
+TEST_CASE("Unit_hipPointerGetAttributes_GpuIter_Managed__Memory") {
+  int deviceCount{0};
+  HIP_CHECK(hipGetDeviceCount(&deviceCount));
+  REQUIRE(deviceCount != 0);
+  int* managed_ptr{nullptr};
+  for (int i = 0; i < deviceCount; i++) {
+    HIP_CHECK(hipSetDevice(i));
+    int managed_memory = 0;
+    HIP_CHECK(hipDeviceGetAttribute(&managed_memory,
+    hipDeviceAttributeManagedMemory, i));
+    if (!managed_memory) {
+      INFO("Managed memory access not supported on the device" << i);
+    } else {
+      // Allocating the memory and test changes only if managed
+      // memory is available
+      HIP_CHECK(hipMallocManaged(&managed_ptr, sizeof(int)));
+      REQUIRE(managed_ptr != nullptr);
+      hipPointerAttribute_t attributes{};
+      HIP_CHECK(hipPointerGetAttributes(&attributes, managed_ptr));
+      REQUIRE(attributes.device == i);  // Device Check
+      REQUIRE(attributes.isManaged);
+      REQUIRE(attributes.type == hipMemoryTypeManaged);
+      REQUIRE(attributes.devicePointer == managed_ptr);
+      HIP_CHECK(hipFree(managed_ptr));
+    }
+  }
+}
+/**
+ * Test Description
+ * ------------------------
+ *  - Unregistered memory type attribute verification.
+ * Test source
+ * ------------------------
+ *  - unit/memory/hipPointerGetAttributes.cc
+ * Test requirements
+ * ------------------------
+ *  - HIP_VERSION >= 6.0
+ */
+TEST_CASE("Unit_hipPointerGetAttributes_GpuIter_Unregistered_Memory") {
+  int deviceCount{0};
+  HIP_CHECK(hipGetDeviceCount(&deviceCount));
+  REQUIRE(deviceCount != 0);
+  int* ptr1{nullptr}; int* ptr2{nullptr}; int* ptr3{nullptr};
+  for (int i = 0; i < deviceCount; i++) {
+    HIP_CHECK(hipSetDevice(i));
+    int N = 5;
+    ptr1 = new int;
+    ptr2 = reinterpret_cast<int*>(malloc(sizeof(int)));
+    ptr3 = reinterpret_cast<int *>(calloc(N, sizeof(int)));
+    int ptr4[1000];
+    #ifdef __linux__
+      void* ptr5{nullptr};
+      ptr5 = mmap ( NULL, N*sizeof(int),
+        PROT_READ | PROT_WRITE, MAP_PRIVATE | MAP_ANONYMOUS, 0, 0);
+
+      if (ptr5 == MAP_FAILED) {
+        INFO("Mapping Failed\n");
+        REQUIRE(false);
+      }
+      REQUIRE(ptr5 != nullptr);
+      hipPointerAttribute_t attributes1{};
+      HIP_CHECK(hipPointerGetAttributes(&attributes1, ptr5));
+      REQUIRE(attributes1.type == hipMemoryTypeUnregistered);
+      int err = munmap(ptr5, N*sizeof(int));
+      if (err != 0) {
+       INFO("UnMapping Failed\n");
+       REQUIRE(false);
+      }
+    #endif
+    REQUIRE(ptr1 != nullptr);
+    REQUIRE(ptr2 != nullptr);
+    REQUIRE(ptr3 != nullptr);
+
+    hipPointerAttribute_t attributes{};
+    HIP_CHECK(hipPointerGetAttributes(&attributes, ptr1));
+    REQUIRE(attributes.type == hipMemoryTypeUnregistered);
+    HIP_CHECK(hipPointerGetAttributes(&attributes, ptr2));
+    REQUIRE(attributes.type == hipMemoryTypeUnregistered);
+    HIP_CHECK(hipPointerGetAttributes(&attributes, ptr3));
+    REQUIRE(attributes.type == hipMemoryTypeUnregistered);
+    HIP_CHECK(hipPointerGetAttributes(&attributes, ptr4));
+    REQUIRE(attributes.type == hipMemoryTypeUnregistered);
+
+    delete ptr1;
+    free(ptr2);
+    free(ptr3);
+  }
+}