[SYCL][NATIVECPU] more ur adapter implementations (#10972)

Adding more UR adapter functionality to pass more tests related to
buffers and subbuffers, and to enable `sycl-ls --verbose`.

Author: uwedolinsky

sycl/plugins/unified_runtime/ur/adapters/native_cpu/common.hpp

@@ -40,3 +40,16 @@ extern thread_local char ErrorMessage[MaxMessageSize];
                 << std::endl;                                                  \
     }                                                                          \
     return UR_RESULT_ERROR_UNSUPPORTED_FEATURE;
+
+// Base class to store common data
+struct _ur_object {
+  ur_shared_mutex Mutex;
+};
+
+struct RefCounted {
+  std::atomic_uint32_t _refCount;
+  void incrementReferenceCount() { _refCount++; }
+  void decrementReferenceCount() { _refCount--; }
+  RefCounted() : _refCount{1} {}
+  uint32_t getReferenceCount() const { return _refCount; }
+};

sycl/plugins/unified_runtime/ur/adapters/native_cpu/device.cpp

@@ -73,10 +73,17 @@ UR_APIEXPORT ur_result_t UR_APICALL urDeviceGetInfo(ur_device_handle_t hDevice,
     return ReturnValue("0.0.0");
   case UR_DEVICE_INFO_VENDOR:
     return ReturnValue("Intel(R) Corporation");
+  case UR_DEVICE_INFO_BACKEND_RUNTIME_VERSION:
+    // TODO : CHECK
+    return ReturnValue("0.0.0");
   case UR_DEVICE_INFO_IMAGE2D_MAX_WIDTH:
     return ReturnValue(size_t{8192});
   case UR_DEVICE_INFO_IMAGE2D_MAX_HEIGHT:
     return ReturnValue(size_t{8192});
+  case UR_DEVICE_INFO_IMAGE_MAX_BUFFER_SIZE:
+    return ReturnValue(size_t(65536 /*todo: min if aspect::image*/));
+  case UR_DEVICE_INFO_MAX_SAMPLERS:
+    return ReturnValue(uint32_t{16 /*todo: min if aspect::image*/});
   case UR_DEVICE_INFO_HOST_UNIFIED_MEMORY:
     return ReturnValue(bool{1});
   case UR_DEVICE_INFO_EXTENSIONS:
@@ -114,6 +121,20 @@ UR_APIEXPORT ur_result_t UR_APICALL urDeviceGetInfo(ur_device_handle_t hDevice,
   case UR_DEVICE_INFO_IMAGE3D_MAX_DEPTH:
     // Default minimum values required by the SYCL specification.
     return ReturnValue(size_t{2048});
+  case UR_DEVICE_INFO_HALF_FP_CONFIG: {
+    // todo:
+    ur_device_fp_capability_flags_t HalfFPValue = 0;
+    return ReturnValue(HalfFPValue);
+  }
+  case UR_DEVICE_INFO_SINGLE_FP_CONFIG: {
+    // todo
+    ur_device_fp_capability_flags_t SingleFPValue = 0;
+    return ReturnValue(SingleFPValue);
+  }
+  case UR_DEVICE_INFO_DOUBLE_FP_CONFIG: {
+    ur_device_fp_capability_flags_t DoubleFPValue = 0;
+    return ReturnValue(DoubleFPValue);
+  }
   case UR_DEVICE_INFO_MAX_WORK_ITEM_DIMENSIONS:
     return ReturnValue(uint32_t{3});
   case UR_DEVICE_INFO_PARTITION_TYPE:
@@ -193,6 +214,9 @@ UR_APIEXPORT ur_result_t UR_APICALL urDeviceGetInfo(ur_device_handle_t hDevice,
   case UR_DEVICE_INFO_GLOBAL_MEM_SIZE:
     // TODO : CHECK
     return ReturnValue(uint64_t{0});
+  case UR_DEVICE_INFO_LOCAL_MEM_SIZE:
+    // TODO : CHECK
+    return ReturnValue(uint64_t{0});
   case UR_DEVICE_INFO_MAX_CONSTANT_BUFFER_SIZE:
     // TODO : CHECK
     return ReturnValue(uint64_t{0});

sycl/plugins/unified_runtime/ur/adapters/native_cpu/kernel.cpp

@@ -69,10 +69,31 @@ UR_APIEXPORT ur_result_t UR_APICALL urKernelGetInfo(ur_kernel_handle_t hKernel,
                                                     size_t *pPropSizeRet) {
   std::ignore = hKernel;
   std::ignore = propName;
-  std::ignore = propSize;
   std::ignore = pPropValue;
-  std::ignore = pPropSizeRet;
 
+  UrReturnHelper ReturnValue(propSize, pPropValue, pPropSizeRet);
+  // todo: check if we need this
+  // std::shared_lock<ur_shared_mutex> Guard(hKernel->Mutex);
+  switch (propName) {
+    //  case UR_KERNEL_INFO_CONTEXT:
+    //    return ReturnValue(ur_context_handle_t{ hKernel->Program->Context });
+    //  case UR_KERNEL_INFO_PROGRAM:
+    //    return ReturnValue(ur_program_handle_t{ Kernel->Program });
+  case UR_KERNEL_INFO_FUNCTION_NAME:
+    if (hKernel->_name) {
+      return ReturnValue(hKernel->_name);
+    }
+    return UR_RESULT_ERROR_INVALID_FUNCTION_NAME;
+    //  case UR_KERNEL_INFO_NUM_ARGS:
+    //    return ReturnValue(uint32_t{ Kernel->ZeKernelProperties->numKernelArgs
+    //    });
+  case UR_KERNEL_INFO_REFERENCE_COUNT:
+    return ReturnValue(uint32_t{hKernel->getReferenceCount()});
+  case UR_KERNEL_INFO_ATTRIBUTES:
+    return ReturnValue("");
+  default:
+    return UR_RESULT_ERROR_INVALID_VALUE;
+  }
   DIE_NO_IMPLEMENTATION
 }
 
@@ -92,7 +113,8 @@ urKernelGetGroupInfo(ur_kernel_handle_t hKernel, ur_device_handle_t hDevice,
     return returnValue(global_work_size, 3);
   }
   case UR_KERNEL_GROUP_INFO_WORK_GROUP_SIZE: {
-    size_t max_threads = 0;
+    // todo: set proper values
+    size_t max_threads = 128;
     return returnValue(max_threads);
   }
   case UR_KERNEL_GROUP_INFO_COMPILE_WORK_GROUP_SIZE: {
@@ -104,7 +126,8 @@ urKernelGetGroupInfo(ur_kernel_handle_t hKernel, ur_device_handle_t hDevice,
     return returnValue(static_cast<uint64_t>(bytes));
   }
   case UR_KERNEL_GROUP_INFO_PREFERRED_WORK_GROUP_SIZE_MULTIPLE: {
-    int warpSize = 0;
+    // todo: set proper values
+    int warpSize = 16;
     return returnValue(static_cast<size_t>(warpSize));
   }
   case UR_KERNEL_GROUP_INFO_PRIVATE_MEM_SIZE: {
@@ -125,17 +148,34 @@ urKernelGetSubGroupInfo(ur_kernel_handle_t hKernel, ur_device_handle_t hDevice,
                         void *pPropValue, size_t *pPropSizeRet) {
   std::ignore = hKernel;
   std::ignore = hDevice;
-  std::ignore = propName;
-  std::ignore = propSize;
-  std::ignore = pPropValue;
-  std::ignore = pPropSizeRet;
 
-  DIE_NO_IMPLEMENTATION
+  UrReturnHelper ReturnValue(propSize, pPropValue, pPropSizeRet);
+  switch (propName) {
+  case UR_KERNEL_SUB_GROUP_INFO_MAX_SUB_GROUP_SIZE: {
+    // todo: set proper values
+    int WarpSize = 8;
+    return ReturnValue(static_cast<uint32_t>(WarpSize));
+  }
+  case UR_KERNEL_SUB_GROUP_INFO_MAX_NUM_SUB_GROUPS: {
+    // todo: set proper values
+    int MaxWarps = 32;
+    return ReturnValue(static_cast<uint32_t>(MaxWarps));
+  }
+  case UR_KERNEL_SUB_GROUP_INFO_COMPILE_NUM_SUB_GROUPS: {
+    // todo: set proper values
+    return ReturnValue(0);
+  }
+  case UR_KERNEL_SUB_GROUP_INFO_SUB_GROUP_SIZE_INTEL: {
+    // todo: set proper values
+    return ReturnValue(0);
+  }
+  }
+  DIE_NO_IMPLEMENTATION;
 }
 
 UR_APIEXPORT ur_result_t UR_APICALL urKernelRetain(ur_kernel_handle_t hKernel) {
-  std::ignore = hKernel;
-  DIE_NO_IMPLEMENTATION
+  hKernel->incrementReferenceCount();
+  return UR_RESULT_SUCCESS;
 }
 
 UR_APIEXPORT ur_result_t UR_APICALL

sycl/plugins/unified_runtime/ur/adapters/native_cpu/kernel.hpp

@@ -8,6 +8,7 @@
 
 #pragma once
 
+#include "common.hpp"
 #include <sycl/detail/native_cpu.hpp>
 #include <ur_api.h>
 
@@ -16,7 +17,7 @@ using nativecpu_kernel_t = void(const sycl::detail::NativeCPUArgDesc *,
 using nativecpu_ptr_t = nativecpu_kernel_t *;
 using nativecpu_task_t = std::function<nativecpu_kernel_t>;
 
-struct ur_kernel_handle_t_ {
+struct ur_kernel_handle_t_ : RefCounted {
 
   ur_kernel_handle_t_(const char *name, nativecpu_task_t subhandler)
       : _name{name}, _subhandler{subhandler} {}

sycl/plugins/unified_runtime/ur/adapters/native_cpu/memory.cpp

@@ -49,11 +49,11 @@ UR_APIEXPORT ur_result_t UR_APICALL urMemBufferCreate(
   ur_mem_handle_t_ *retMem;
 
   if (useHostPtr) {
-    retMem = new ur_mem_handle_t_(pProperties->pHost);
+    retMem = new _ur_buffer(hContext, pProperties->pHost);
   } else if (copyHostPtr) {
-    retMem = new ur_mem_handle_t_(pProperties->pHost, size);
+    retMem = new _ur_buffer(hContext, pProperties->pHost, size);
   } else {
-    retMem = new ur_mem_handle_t_(size);
+    retMem = new _ur_buffer(hContext, size);
   }
 
   *phBuffer = retMem;
@@ -82,13 +82,30 @@ UR_APIEXPORT ur_result_t UR_APICALL urMemBufferPartition(
     ur_mem_handle_t hBuffer, ur_mem_flags_t flags,
     ur_buffer_create_type_t bufferCreateType, const ur_buffer_region_t *pRegion,
     ur_mem_handle_t *phMem) {
-  std::ignore = hBuffer;
-  std::ignore = flags;
+
   std::ignore = bufferCreateType;
-  std::ignore = pRegion;
-  std::ignore = phMem;
+  UR_ASSERT(hBuffer && !hBuffer->isImage() &&
+                !(static_cast<_ur_buffer *>(hBuffer))->isSubBuffer(),
+            UR_RESULT_ERROR_INVALID_MEM_OBJECT);
 
-  DIE_NO_IMPLEMENTATION
+  std::shared_lock<ur_shared_mutex> Guard(hBuffer->Mutex);
+
+  if (flags != UR_MEM_FLAG_READ_WRITE) {
+    die("urMemBufferPartition: NativeCPU implements only read-write buffer,"
+        "no read-only or write-only yet.");
+  }
+
+  try {
+    auto partitionedBuffer = new _ur_buffer(static_cast<_ur_buffer *>(hBuffer),
+                                            pRegion->origin, pRegion->size);
+    *phMem = reinterpret_cast<ur_mem_handle_t>(partitionedBuffer);
+  } catch (const std::bad_alloc &) {
+    return UR_RESULT_ERROR_OUT_OF_HOST_MEMORY;
+  } catch (...) {
+    return UR_RESULT_ERROR_UNKNOWN;
+  }
+
+  return UR_RESULT_SUCCESS;
 }
 
 UR_APIEXPORT ur_result_t UR_APICALL

sycl/plugins/unified_runtime/ur/adapters/native_cpu/memory.hpp

@@ -12,17 +12,23 @@
 #include <cstdlib>
 #include <cstring>
 
-struct ur_mem_handle_t_ {
-  ur_mem_handle_t_(size_t Size)
-      : _mem{static_cast<char *>(malloc(Size))}, _ownsMem{true}, _refCount{1} {}
+#include "common.hpp"
+#include "context.hpp"
 
-  ur_mem_handle_t_(void *HostPtr, size_t Size)
-      : _mem{static_cast<char *>(malloc(Size))}, _ownsMem{true}, _refCount{1} {
+struct ur_mem_handle_t_ : _ur_object {
+  ur_mem_handle_t_(size_t Size, bool _IsImage)
+      : _mem{static_cast<char *>(malloc(Size))}, _ownsMem{true},
+        IsImage{_IsImage} {}
+
+  ur_mem_handle_t_(void *HostPtr, size_t Size, bool _IsImage)
+      : _mem{static_cast<char *>(malloc(Size))}, _ownsMem{true},
+        IsImage{_IsImage} {
     memcpy(_mem, HostPtr, Size);
   }
 
-  ur_mem_handle_t_(void *HostPtr)
-      : _mem{static_cast<char *>(HostPtr)}, _ownsMem{false}, _refCount{1} {}
+  ur_mem_handle_t_(void *HostPtr, bool _IsImage)
+      : _mem{static_cast<char *>(HostPtr)}, _ownsMem{false}, IsImage{_IsImage} {
+  }
 
   ~ur_mem_handle_t_() {
     if (_ownsMem) {
@@ -32,7 +38,36 @@ struct ur_mem_handle_t_ {
 
   void decrementRefCount() noexcept { _refCount--; }
 
+  // Method to get type of the derived object (image or buffer)
+  bool isImage() const { return this->IsImage; }
+
   char *_mem;
   bool _ownsMem;
-  std::atomic_uint32_t _refCount;
+  std::atomic_uint32_t _refCount = {1};
+
+private:
+  const bool IsImage;
+};
+
+struct _ur_buffer final : ur_mem_handle_t_ {
+  // Buffer constructor
+  _ur_buffer(ur_context_handle_t /* Context*/, void *HostPtr)
+      : ur_mem_handle_t_(HostPtr, false) {}
+  _ur_buffer(ur_context_handle_t /* Context*/, void *HostPtr, size_t Size)
+      : ur_mem_handle_t_(HostPtr, Size, false) {}
+  _ur_buffer(ur_context_handle_t /* Context*/, size_t Size)
+      : ur_mem_handle_t_(Size, false) {}
+  _ur_buffer(_ur_buffer *b, size_t Offset, size_t Size)
+      : ur_mem_handle_t_(b->_mem + Offset, false), SubBuffer(b) {
+    SubBuffer.Origin = Offset;
+  }
+
+  bool isSubBuffer() const { return SubBuffer.Parent != nullptr; }
+
+  struct BB {
+    BB(_ur_buffer *b) : Parent(b) {}
+    BB() : BB(nullptr) {}
+    _ur_buffer *const Parent;
+    size_t Origin; // only valid if Parent != nullptr
+  } SubBuffer;
 };

sycl/plugins/unified_runtime/ur/adapters/native_cpu/program.cpp

@@ -87,9 +87,8 @@ urProgramLink(ur_context_handle_t hContext, uint32_t count,
 
 UR_APIEXPORT ur_result_t UR_APICALL
 urProgramRetain(ur_program_handle_t hProgram) {
-  std::ignore = hProgram;
-
-  DIE_NO_IMPLEMENTATION
+  hProgram->incrementReferenceCount();
+  return UR_RESULT_SUCCESS;
 }
 
 UR_APIEXPORT ur_result_t UR_APICALL

sycl/plugins/unified_runtime/ur/adapters/native_cpu/program.hpp

@@ -13,16 +13,14 @@
 #include "context.hpp"
 #include <map>
 
-struct ur_program_handle_t_ {
+struct ur_program_handle_t_ : RefCounted {
   ur_program_handle_t_(ur_context_handle_t ctx, const unsigned char *pBinary)
-      : _ctx{ctx}, _ptr{pBinary}, _refCount{1} {}
+      : _ctx{ctx}, _ptr{pBinary} {}
 
   uint32_t getReferenceCount() const noexcept { return _refCount; }
 
   ur_context_handle_t _ctx;
   const unsigned char *_ptr;
-  std::atomic_uint32_t _refCount;
-
   struct _compare {
     bool operator()(char const *a, char const *b) const {
       return std::strcmp(a, b) < 0;

sycl/plugins/unified_runtime/ur/adapters/native_cpu/usm.cpp

@@ -52,10 +52,14 @@ urUSMSharedAlloc(ur_context_handle_t hContext, ur_device_handle_t hDevice,
   std::ignore = hDevice;
   std::ignore = pUSMDesc;
   std::ignore = pool;
-  std::ignore = size;
-  std::ignore = ppMem;
 
-  DIE_NO_IMPLEMENTATION;
+  UR_ASSERT(ppMem, UR_RESULT_ERROR_INVALID_NULL_POINTER);
+  // TODO: Check Max size when UR_DEVICE_INFO_MAX_MEM_ALLOC_SIZE is implemented
+  UR_ASSERT(size > 0, UR_RESULT_ERROR_INVALID_USM_SIZE);
+
+  *ppMem = malloc(size);
+
+  return UR_RESULT_SUCCESS;
 }
 
 UR_APIEXPORT ur_result_t UR_APICALL urUSMFree(ur_context_handle_t hContext,