Address review feedback

Author: Georgi Mirazchiyski

source/adapters/hip/device.cpp

@@ -880,27 +880,26 @@ UR_APIEXPORT ur_result_t UR_APICALL urDeviceGetInfo(ur_device_handle_t hDevice,
     return ReturnValue(ur_bool_t{false});
   }
   case UR_DEVICE_INFO_BINDLESS_SAMPLED_IMAGE_FETCH_1D_USM_EXP: {
-    // HIP does support fetching 1D USM sampled image data.
+    // HIP supports fetching 1D USM sampled image data.
     // TODO: DPC++ doesn't implement the required builtins for SYCL.
     return ReturnValue(ur_bool_t{false});
   }
   case UR_DEVICE_INFO_BINDLESS_SAMPLED_IMAGE_FETCH_1D_EXP: {
     // HIP does not support fetching 1D non-USM sampled image data.
-    // TODO: DPC++ doesn't implement the required builtins for SYCL.
     return ReturnValue(ur_bool_t{false});
   }
   case UR_DEVICE_INFO_BINDLESS_SAMPLED_IMAGE_FETCH_2D_USM_EXP: {
-    // HIP does support fetching 2D USM sampled image data.
+    // HIP supports fetching 2D USM sampled image data.
     // TODO: DPC++ doesn't implement the required builtins for SYCL.
     return ReturnValue(ur_bool_t{false});
   }
   case UR_DEVICE_INFO_BINDLESS_SAMPLED_IMAGE_FETCH_2D_EXP: {
-    // HIP does support fetching 2D non-USM sampled image data.
+    // HIP supports fetching 2D non-USM sampled image data.
     // TODO: DPC++ doesn't implement the required builtins for SYCL.
     return ReturnValue(ur_bool_t{false});
   }
   case UR_DEVICE_INFO_BINDLESS_SAMPLED_IMAGE_FETCH_3D_EXP: {
-    // HIP does support fetching 3D non-USM sampled image data.
+    // HIP supports fetching 3D non-USM sampled image data.
     // TODO: DPC++ doesn't implement the required builtins for SYCL.
     return ReturnValue(ur_bool_t{false});
   }
@@ -916,12 +915,12 @@ UR_APIEXPORT ur_result_t UR_APICALL urDeviceGetInfo(ur_device_handle_t hDevice,
     return ReturnValue(ur_bool_t{false});
   }
   case UR_DEVICE_INFO_BINDLESS_SAMPLE_1D_USM_EXP: {
-    // HIP does support sampling 1D USM sampled image data.
+    // HIP supports sampling 1D USM sampled image data.
     return ReturnValue(
         static_cast<ur_bool_t>(hDevice->supportsHardwareImages()));
   }
   case UR_DEVICE_INFO_BINDLESS_SAMPLE_2D_USM_EXP: {
-    // HIP does support sampling 2D USM sampled image data.
+    // HIP supports sampling 2D USM sampled image data.
     return ReturnValue(
         static_cast<ur_bool_t>(hDevice->supportsHardwareImages()));
   }

source/adapters/hip/device.hpp

@@ -13,8 +13,6 @@
 
 #include <ur/ur.hpp>
 
-#include <map>
-
 /// UR device mapping to a hipDevice_t.
 /// Includes an observer pointer to the platform,
 /// and implements the reference counting semantics since
@@ -36,7 +34,7 @@ struct ur_device_handle_t_ {
   int DeviceMaxLocalMem{0};
   int ManagedMemSupport{0};
   int ConcurrentManagedAccess{0};
-  int HardwareImageSupport{0};
+  bool HardwareImageSupport{false};
 
 public:
   ur_device_handle_t_(native_type HipDevice, hipEvent_t EvBase,
@@ -61,9 +59,11 @@ struct ur_device_handle_t_ {
         &ConcurrentManagedAccess, hipDeviceAttributeConcurrentManagedAccess,
         HIPDevice));
     // Check if texture functions are supported in the HIP host runtime.
-    UR_CHECK_ERROR(hipDeviceGetAttribute(
-        &HardwareImageSupport, hipDeviceAttributeImageSupport, HIPDevice));
-    detail::ur::assertion(HardwareImageSupport >= 0);
+    int Ret{};
+    UR_CHECK_ERROR(
+        hipDeviceGetAttribute(&Ret, hipDeviceAttributeImageSupport, HIPDevice));
+    detail::ur::assertion(Ret == 0 || Ret == 1);
+    HardwareImageSupport = Ret == 1;
   }
 
   ~ur_device_handle_t_() noexcept(false) {}
@@ -96,12 +96,7 @@ struct ur_device_handle_t_ {
     return ConcurrentManagedAccess;
   };
 
-  bool supportsHardwareImages() const noexcept {
-    return HardwareImageSupport ? true : false;
-  }
-
-  // Used for bookkeeping for mipmapped array leaks in mapping external memory.
-  std::map<hipArray_t, hipMipmappedArray_t> ChildHipArrayFromMipmapMap;
+  bool supportsHardwareImages() const noexcept { return HardwareImageSupport; }
 };
 
 int getAttribute(ur_device_handle_t Device, hipDeviceAttribute_t Attribute);

source/adapters/hip/image.cpp

@@ -128,6 +128,7 @@ hipToUrImageChannelFormat(hipArray_Format hip_format,
     *return_image_channel_type = TO;                                           \
     return UR_RESULT_SUCCESS;                                                  \
   }
+
     HIP_TO_UR_IMAGE_CHANNEL_TYPE(HIP_AD_FORMAT_UNSIGNED_INT8,
                                  UR_IMAGE_CHANNEL_TYPE_UNSIGNED_INT8);
     HIP_TO_UR_IMAGE_CHANNEL_TYPE(HIP_AD_FORMAT_UNSIGNED_INT16,
@@ -144,6 +145,8 @@ hipToUrImageChannelFormat(hipArray_Format hip_format,
                                  UR_IMAGE_CHANNEL_TYPE_HALF_FLOAT);
     HIP_TO_UR_IMAGE_CHANNEL_TYPE(HIP_AD_FORMAT_FLOAT,
                                  UR_IMAGE_CHANNEL_TYPE_FLOAT);
+
+#undef HIP_TO_UR_IMAGE_CHANNEL_TYPE
   default:
     return UR_RESULT_ERROR_UNSUPPORTED_IMAGE_FORMAT;
   }
@@ -156,7 +159,6 @@ ur_result_t urTextureCreate(ur_sampler_handle_t hSampler,
                             ur_exp_image_native_handle_t *phRetImage) {
 
   try {
-    /// pi_sampler_properties
     /// Layout of UR samplers for HIP
     ///
     /// Sampler property layout:
@@ -425,12 +427,6 @@ UR_APIEXPORT ur_result_t UR_APICALL urBindlessImagesImageFreeExp(
   try {
     hipArray_t ImageArray = reinterpret_cast<hipArray_t>(hImageMem);
     UR_CHECK_ERROR(hipArrayDestroy(ImageArray));
-    if (auto It = hDevice->ChildHipArrayFromMipmapMap.find(ImageArray);
-        It != hDevice->ChildHipArrayFromMipmapMap.end()) {
-      UR_CHECK_ERROR(hipMipmappedArrayDestroy(
-          static_cast<hipMipmappedArray_t>(It->second)));
-      hDevice->ChildHipArrayFromMipmapMap.erase(It);
-    }
   } catch (ur_result_t Err) {
     return Err;
   } catch (...) {
@@ -625,7 +621,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urBindlessImagesImageCopyExp(
 
         if (memType == hipMemoryTypeArray) {
           // HIP doesn not provide async copies between host and image arrays
-          // memory in versions early than 6.2.
+          // memory in versions earlier than 6.2.
 #if HIP_VERSION >= 60200000
           UR_CHECK_ERROR(
               hipMemcpyHtoAAsync(static_cast<hipArray_t>(pDst),
@@ -733,7 +729,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urBindlessImagesImageCopyExp(
 
         if (memType == hipMemoryTypeArray) {
           // HIP doesn not provide async copies between image arrays and host
-          // memory in versions early than 6.2.
+          // memory in versions earlier than 6.2.
 #if HIP_VERSION >= 60200000
           UR_CHECK_ERROR(hipMemcpyAtoHAsync(
               DstWithOffset, static_cast<hipArray_t>(const_cast<void *>(pSrc)),
@@ -1095,8 +1091,9 @@ UR_APIEXPORT ur_result_t UR_APICALL urBindlessImagesImportExternalMemoryExp(
           extMemDesc.type = hipExternalMemoryHandleTypeOpaqueWin32;
           break;
         case UR_EXP_EXTERNAL_MEM_TYPE_WIN32_NT_DX12_RESOURCE:
-          // Memory descriptor flag values such as hipExternalMemoryDedicatedare
-          // not available before HIP 5.6, so we safely fallback to unsupported.
+          // Memory descriptor flag values such as hipExternalMemoryDedicated
+          // are not available before HIP 5.6, so we safely fallback to marking
+          // this as an unsupported.
 #if HIP_VERSION >= 50600000
           extMemDesc.type = hipExternalMemoryHandleTypeD3D12Resource;
           extMemDesc.flags = hipExternalMemoryDedicated;
@@ -1230,6 +1227,7 @@ UR_APIEXPORT ur_result_t UR_APICALL urBindlessImagesImportExternalSemaphoreExp(
           extSemDesc.type = hipExternalSemaphoreHandleTypeD3D12Fence;
           break;
         case UR_EXP_EXTERNAL_SEMAPHORE_TYPE_OPAQUE_FD:
+          [[fallthrough]];
         default:
           return UR_RESULT_ERROR_INVALID_VALUE;
         }

source/adapters/hip/sampler.hpp

@@ -55,6 +55,8 @@ struct ur_sampler_handle_t_ {
   }
 
   ur_sampler_addressing_mode_t getAddressingModeDim(size_t i) const noexcept {
+    // valid dimensions are 0,1,2
+    assert(i < 3);
     return static_cast<ur_sampler_addressing_mode_t>((Props >> (2 + (i * 3))) &
                                                      0b111);
   }