185a23e865
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@intel.com>
77 lines
2.4 KiB
C++
77 lines
2.4 KiB
C++
//===- llvm/unittest/Frontend/PropertySetRegistry.cpp ---------------------===//
|
|
//
|
|
// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
|
|
// See https://llvm.org/LICENSE.txt for license information.
|
|
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
|
|
//
|
|
//===----------------------------------------------------------------------===//
|
|
|
|
#include "llvm/ADT/SmallVector.h"
|
|
#include "llvm/Frontend/Offloading/PropertySet.h"
|
|
#include "llvm/Support/MemoryBuffer.h"
|
|
#include "gtest/gtest.h"
|
|
|
|
using namespace llvm::offloading;
|
|
using namespace llvm;
|
|
|
|
void checkSerialization(const PropertySetRegistry &PSR) {
|
|
SmallString<0> Serialized;
|
|
raw_svector_ostream OS(Serialized);
|
|
writePropertiesToJSON(PSR, OS);
|
|
auto PSR2 = readPropertiesFromJSON({Serialized, ""});
|
|
ASSERT_EQ("", toString(PSR2.takeError()));
|
|
EXPECT_EQ(PSR, *PSR2);
|
|
}
|
|
|
|
TEST(PropertySetRegistryTest, PropertySetRegistry) {
|
|
PropertySetRegistry PSR;
|
|
checkSerialization(PSR);
|
|
|
|
PSR["Category1"]["Prop1"] = 42U;
|
|
PSR["Category1"]["Prop2"] = ByteArray(StringRef("Hello").bytes());
|
|
PSR["Category2"]["A"] = ByteArray{0, 4, 16, 32, 255};
|
|
checkSerialization(PSR);
|
|
|
|
PSR = PropertySetRegistry();
|
|
PSR["ABC"]["empty_array"] = ByteArray();
|
|
PSR["ABC"]["max_val"] = std::numeric_limits<uint32_t>::max();
|
|
checkSerialization(PSR);
|
|
}
|
|
|
|
TEST(PropertySetRegistryTest, IllFormedJSON) {
|
|
SmallString<0> Input;
|
|
|
|
// Invalid json
|
|
Input = "{ invalid }";
|
|
auto Res = readPropertiesFromJSON({Input, ""});
|
|
EXPECT_NE("", toString(Res.takeError()));
|
|
|
|
Input = "";
|
|
Res = readPropertiesFromJSON({Input, ""});
|
|
EXPECT_NE("", toString(Res.takeError()));
|
|
|
|
// Not a JSON object
|
|
Input = "[1, 2, 3]";
|
|
Res = readPropertiesFromJSON({Input, ""});
|
|
EXPECT_NE("", toString(Res.takeError()));
|
|
|
|
// Property set not an object
|
|
Input = R"({ "Category": 42 })";
|
|
Res = readPropertiesFromJSON({Input, ""});
|
|
EXPECT_NE("", toString(Res.takeError()));
|
|
|
|
// Property value has non string/non-integer type
|
|
Input = R"({ "Category": { "Prop": [1, 2, 3] } })";
|
|
Res = readPropertiesFromJSON({Input, ""});
|
|
EXPECT_NE("", toString(Res.takeError()));
|
|
|
|
// Property value is an invalid base64 string
|
|
Input = R"({ "Category": { "Prop": ";" } })";
|
|
Res = readPropertiesFromJSON({Input, ""});
|
|
EXPECT_NE("", toString(Res.takeError()));
|
|
|
|
Input = R"({ "Category": { "Prop": "!@#$" } })";
|
|
Res = readPropertiesFromJSON({Input, ""});
|
|
EXPECT_NE("", toString(Res.takeError()));
|
|
}
|