[all-commits] [llvm/llvm-project] 185a23: [SYCL] Add property set types and JSON representat...

[Justin Cai via All-commits]

  Branch: refs/heads/main
  Home:   https://github.com/llvm/llvm-project
  Commit: 185a23e865c31145588276e7e0db2460e01eb703
      https://github.com/llvm/llvm-project/commit/185a23e865c31145588276e7e0db2460e01eb703
  Author: Justin Cai <justin.cai at intel.com>
  Date:   2025-08-01 (Fri, 01 Aug 2025)

  Changed paths:
    A llvm/include/llvm/Frontend/Offloading/PropertySet.h
    M llvm/lib/Frontend/Offloading/CMakeLists.txt
    A llvm/lib/Frontend/Offloading/PropertySet.cpp
    M llvm/unittests/Frontend/CMakeLists.txt
    A llvm/unittests/Frontend/PropertySetRegistryTest.cpp

  Log Message:
  -----------
  [SYCL] Add property set types and JSON representation (#147321)

This PR adds the `PropertySet` type, along with a pair of functions used
to serialize and deserialize into a JSON representation. A property set
is a key-value map, with values being one of 2 types - uint32 or byte
array. A property set registry is a collection of property sets, indexed
by a "category" name.

In SYCL offloading, property sets will be used to communicate metadata
about device images needed by the SYCL runtime. For example, there is a
property set which has a byte array containing the numeric ID, offset,
and size of each SYCL2020 spec constant. Another example is a property
set describing the optional kernel features used in the module: does it
use fp64? fp16? atomic64?

This metadata will be computed by `clang-sycl-linker` and the JSON
representation will be inserted in the string table of each
output `OffloadBinary`. This JSON will be consumed the SYCL offload
wrapper and will be lowered to the binary form SYCL runtime expects.

For example, consider this SYCL program that calls a kernel that uses
fp64:

```c++
#include <sycl/sycl.hpp>

using namespace sycl;
class MyKernel;

int main() {
  queue q;
  auto *p = malloc_shared<double>(1, q);
  *p = .1;
  q.single_task<MyKernel>([=]{ *p *= 2; }).wait();
  std::cout << *p << "\n";
  free(p, q);
}
```

The device code for this program would have the kernel marked with
`!sycl_used_aspects`:

```
define spir_kernel void @_ZTS8MyKernel([...]) !sycl_used_aspects !n { [...] }
!n = {i32 6}
```

`clang-sycl-linker` would recognize this metadata and then would output
the following JSON in the `OffloadBinary`'s key-value map:

```
{
  "SYCL/device requirements": {
    // aspects contains a list of sycl::aspect values used
    // by the module; in this case just the value 6 encoded 
    // as a 4-byte little-endian integer 
    "aspects": "BjAwMA=="
  }
}
```

The SYCL offload wrapper would lower those property sets to something
like this:

```c++
struct _sycl_device_binary_property_set_struct {
  char *CategoryName;
  _sycl_device_binary_property *PropertiesBegin;
  _sycl_device_binary_property *PropertiesEnd;
};

struct _sycl_device_binary_property_struct {
  char *PropertyName;  
  void *ValAddr;     
  uint64_t ValSize; 
};

//  
_sycl_device_binary_property_struct device_requirements[] = {
  /* PropertyName */ "aspects",
  /* ValAddr */ [pointer to the bytes 0x06 0x00 0x00 0x00],
  /* ValSize */ 4,
};

_sycl_device_binary_property_set_struct properties[] = {
  /* CategoryName */ "SYCL/device requirements",
  /* PropertiesBegin */ device_requirements,
  /* PropertiesEnd */ std::end(device_requirments),
}
```

---------

Co-authored-by: Arvind Sudarsanam <arvind.sudarsanam at intel.com>



To unsubscribe from these emails, change your notification settings at https://github.com/llvm/llvm-project/settings/notifications