From 12e65466401cf8d8ce13941a85a621ea25b0199e Mon Sep 17 00:00:00 2001
From: =?UTF-8?q?F=C3=A1bio?= <fabio.m.mestre@gmail.com>
Date: Wed, 3 Aug 2022 12:20:07 +0100
Subject: [PATCH] EXSWCPHIPT-138: Update hipHostRegister flag testing (#2835)

[ROCm/hip-tests commit: 2f26a24a5decb4c8455da98678bd49679c17fc3e]
---
 .../catch/unit/memory/hipHostRegister.cc      | 43 ++++++++++++++++---
 1 file changed, 36 insertions(+), 7 deletions(-)

diff --git a/projects/hip-tests/catch/unit/memory/hipHostRegister.cc b/projects/hip-tests/catch/unit/memory/hipHostRegister.cc
index 417ab0857a..3ccc79d834 100644
--- a/projects/hip-tests/catch/unit/memory/hipHostRegister.cc
+++ b/projects/hip-tests/catch/unit/memory/hipHostRegister.cc
@@ -29,6 +29,7 @@ This testfile verifies the following scenarios of hipHostRegister API
 
 #include <hip_test_common.hh>
 #include <hip_test_helper.hh>
+#include "hip/hip_runtime_api.h"
 
 #define OFFSET 128
 static constexpr auto LEN{1024 * 1024};
@@ -160,6 +161,41 @@ template <typename T> __global__ void fill_kernel(T* dataPtr, T value) {
   dataPtr[tid] = value;
 }
 
+TEMPLATE_TEST_CASE("Unit_hipHostRegister_Flags", "", int, float, double) {
+#if HT_AMD
+  HipTest::HIP_SKIP_TEST("EXSWCPHIPT-138");
+  return;
+#else
+
+  size_t sizeBytes = 1 * sizeof(TestType);
+  TestType* hostPtr = reinterpret_cast<TestType*>(malloc(sizeBytes));
+
+  /* Flags aren't used for AMD devices currently */
+  struct FlagType {
+    unsigned int value;
+    bool valid;
+  };
+
+  /* EXSWCPHIPT-29 - 0x08 is hipHostRegisterReadOnly which currently doesn't have a definition in the headers */
+  /* hipHostRegisterIoMemory is a valid flag but requires access to I/O mapped memory to be tested */
+  FlagType flags = GENERATE(
+      FlagType{hipHostRegisterDefault, true}, FlagType{hipHostRegisterPortable, true},
+      FlagType{0x08, true}, FlagType{hipHostRegisterPortable | hipHostRegisterMapped, true},
+      FlagType{hipHostRegisterPortable | hipHostRegisterMapped | 0x08, true}, FlagType{0xF0, false},
+      FlagType{0xFFF2, false}, FlagType{0xFFFFFFFF, false});
+
+  INFO("Testing hipHostRegister flag: " << flags.value);
+  if (flags.valid) {
+    HIP_CHECK(hipHostRegister(hostPtr, sizeBytes, flags.value));
+    HIP_CHECK(hipHostUnregister(hostPtr));
+  } else {
+    HIP_CHECK_ERROR(hipHostRegister(hostPtr, sizeBytes, flags.value), hipErrorInvalidValue);
+  }
+
+  free(hostPtr);
+#endif
+}
+
 TEMPLATE_TEST_CASE("Unit_hipHostRegister_Negative", "", int, float, double) {
   TestType* hostPtr = nullptr;
 
@@ -173,13 +209,6 @@ TEMPLATE_TEST_CASE("Unit_hipHostRegister_Negative", "", int, float, double) {
     HIP_CHECK_ERROR(hipHostRegister(hostPtr, 0, 0), hipErrorInvalidValue);
   }
 
-#if HT_NVIDIA
-  // Flags aren't used for AMD devices currently
-  SECTION("hipHostRegister Negative Test - invalid flag") {
-    HIP_CHECK_ERROR(hipHostRegister(hostPtr, sizeBytes, 0b11111111), hipErrorInvalidValue);
-  }
-#endif
-
   size_t devMemAvail{0}, devMemFree{0};
   HIP_CHECK(hipMemGetInfo(&devMemFree, &devMemAvail));
   auto hostMemFree = HipTest::getMemoryAmount() /* In MB */ * 1024 * 1024;  // In bytes