Skip to content

Commit

Permalink
[SYCL][COMPAT] Add Image Max dims to device_info. Updated Max ND Rang…
Browse files Browse the repository at this point in the history 
…e Size (#13973)

-  `_image1d_max`, `_image2d_max`, and `_image3d_max` member attributes
- Updates the `_max_nd_range_size` implementation conditionally set the
value based on the availability of SYCL_EXT_ONEAPI_MAX_WORK_GROUP_QUERY.
- Updated documentation accordingly.

Minor fix:
- Added missing new line to the output from `test_not_enough_devices`.

---------

Signed-off-by: Alberto Cabrera <[email protected]>
  • Loading branch information
@Alcpz
Alcpz authored Jun 4, 2024
1 parent 4d81735 commit b49303c
Show file tree
Hide file tree
Showing 3 changed files with 150 additions and 6 deletions.
13 changes: 13 additions & 0 deletions sycl/doc/syclcompat/README.md
View file
Original file line number Diff line number Diff line change
Expand Up @@ -54,6 +54,9 @@ If available, the following extensions extend SYCLcompat functionality:

* [sycl_ext_intel_device_info](https://github.com/intel/llvm/blob/sycl/sycl/doc/extensions/supported/sycl_ext_intel_device_info.md) \[Optional\]
* [sycl_ext_oneapi_bfloat16_math_functions](../extensions/experimental/sycl_ext_oneapi_bfloat16_math_functions.asciidoc) \[Optional\]
* [sycl_ext_oneapi_max_work_group_query](
https://github.com/intel/llvm/blob/sycl/sycl/doc/extensions/experimental/sycl_ext_oneapi_max_work_group_query.md)
\[Optional\]

## Usage

Expand Down Expand Up @@ -755,6 +758,9 @@ public:
uint32_t get_device_id() const;
std::array<unsigned char, 16> get_uuid() const;
unsigned int get_global_mem_cache_size() const;
int get_image1d_max() const;
auto get_image2d_max() const;
auto get_image3d_max() const;

void set_name(const char *name);
void set_max_work_item_sizes(const sycl::range<3> max_work_item_sizes);
Expand All @@ -773,13 +779,20 @@ public:
void
set_max_work_items_per_compute_unit(int max_work_items_per_compute_unit);
void set_max_nd_range_size(int max_nd_range_size[]);
void set_max_nd_range_size(sycl::id<3> max_nd_range_size);
void set_memory_clock_rate(unsigned int memory_clock_rate);
void set_memory_bus_width(unsigned int memory_bus_width);
void
set_max_register_size_per_work_group(int max_register_size_per_work_group);
void set_device_id(uint32_t device_id);
void set_uuid(std::array<unsigned char, 16> uuid);
void set_global_mem_cache_size(unsigned int global_mem_cache_size);
void set_image1d_max(size_t image_max_buffer_size);
void set_image2d_max(size_t image_max_width_buffer_size,
size_t image_max_height_buffer_size);
void set_image3d_max(size_t image_max_width_buffer_size,
size_t image_max_height_buffer_size,
size_t image_max_depth_buffer_size);
};
```
Expand Down
49 changes: 49 additions & 0 deletions sycl/include/syclcompat/device.hpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -161,6 +161,9 @@ class device_info {
unsigned int get_global_mem_cache_size() const {
return _global_mem_cache_size;
}
int get_image1d_max() const { return _image1d_max; }
auto get_image2d_max() const { return _image2d_max; }
auto get_image3d_max() const { return _image3d_max; }

// set interface
void set_name(const char *name) {
Expand Down Expand Up @@ -216,6 +219,12 @@ class device_info {
_max_nd_range_size_i[i] = max_nd_range_size[i];
}
}
void set_max_nd_range_size(sycl::id<3> max_nd_range_size) {
for (int i = 0; i < 3; i++) {
_max_nd_range_size[i] = max_nd_range_size[i];
_max_nd_range_size_i[i] = max_nd_range_size[i];
}
}
void set_memory_clock_rate(unsigned int memory_clock_rate) {
_memory_clock_rate = memory_clock_rate;
}
Expand All @@ -231,6 +240,21 @@ class device_info {
void set_global_mem_cache_size(unsigned int global_mem_cache_size) {
_global_mem_cache_size = global_mem_cache_size;
}
void set_image1d_max(size_t image_max_buffer_size) {
_image1d_max = image_max_buffer_size;
}
void set_image2d_max(size_t image_max_width_buffer_size,
size_t image_max_height_buffer_size) {
_image2d_max[0] = image_max_width_buffer_size;
_image2d_max[1] = image_max_height_buffer_size;
}
void set_image3d_max(size_t image_max_width_buffer_size,
size_t image_max_height_buffer_size,
size_t image_max_depth_buffer_size) {
_image3d_max[0] = image_max_width_buffer_size;
_image3d_max[1] = image_max_height_buffer_size;
_image3d_max[2] = image_max_depth_buffer_size;
}

private:
constexpr static size_t NAME_BUFFER_SIZE = 256;
Expand Down Expand Up @@ -259,6 +283,9 @@ class device_info {
int _max_nd_range_size_i[3];
uint32_t _device_id;
std::array<unsigned char, 16> _uuid;
int _image1d_max;
int _image2d_max[2];
int _image3d_max[3];
};

/// device extension
Expand Down Expand Up @@ -370,6 +397,7 @@ class device_ext : public sycl::device {
// by an int
get_info<sycl::info::device::max_work_item_sizes<3>>());
#endif
prop.set_host_unified_memory(has(sycl::aspect::usm_host_allocations));

prop.set_max_clock_frequency(
get_info<sycl::info::device::max_clock_frequency>());
Expand Down Expand Up @@ -422,15 +450,36 @@ Use 64 bits as memory_bus_width default value."
prop.set_max_work_items_per_compute_unit(
get_info<sycl::info::device::max_work_group_size>());
#ifdef SYCL_EXT_ONEAPI_MAX_WORK_GROUP_QUERY
prop.set_max_nd_range_size(
get_info<sycl::ext::oneapi::experimental::info::device::max_work_groups<
3>>());
#else
#if defined(_MSC_VER) && !defined(__clang__)
#pragma message("get_device_info: querying the maximum number \
of work groups is not supported.")
#else
#warning "get_device_info: querying the maximum number of \
work groups is not supported."
#endif
int max_nd_range_size[] = {0x7FFFFFFF, 0x7FFFFFFF, 0x7FFFFFFF};
prop.set_max_nd_range_size(max_nd_range_size);
#endif
// Estimates max register size per work group, feel free to update the value
// according to device properties.
prop.set_max_register_size_per_work_group(65536);
prop.set_global_mem_cache_size(
get_info<sycl::info::device::global_mem_cache_size>());
prop.set_image1d_max(get_info<sycl::info::device::image_max_buffer_size>());
prop.set_image1d_max(get_info<sycl::info::device::image_max_buffer_size>());
prop.set_image2d_max(get_info<sycl::info::device::image2d_max_width>(),
get_info<sycl::info::device::image2d_max_height>());
prop.set_image3d_max(get_info<sycl::info::device::image3d_max_width>(),
get_info<sycl::info::device::image3d_max_height>(),
get_info<sycl::info::device::image3d_max_height>());
out = prop;
}
Expand Down
94 changes: 88 additions & 6 deletions sycl/test-e2e/syclcompat/device/device.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -70,7 +70,7 @@ void test_not_enough_devices() {
try {
syclcompat::select_device(dtf.get_n_devices());
} catch (std::runtime_error const &e) {
std::cout << "Expected SYCL exception caught: " << e.what();
std::cout << "Expected SYCL exception caught: " << e.what() << std::endl;
}
}

Expand Down Expand Up @@ -194,8 +194,6 @@ void test_device_info_api() {
Info.set_max_work_group_size(32);
Info.set_max_sub_group_size(16);
Info.set_max_work_items_per_compute_unit(16);
int SizeArray[3] = {1, 2, 3};
Info.set_max_nd_range_size(SizeArray);

Info.set_host_unified_memory(true);
Info.set_memory_clock_rate(1000);
Expand All @@ -213,9 +211,6 @@ void test_device_info_api() {
assert(Info.get_max_work_group_size() == 32);
assert(Info.get_max_sub_group_size() == 16);
assert(Info.get_max_work_items_per_compute_unit() == 16);
assert(Info.get_max_nd_range_size()[0] == SizeArray[0]);
assert(Info.get_max_nd_range_size()[1] == SizeArray[1]);
assert(Info.get_max_nd_range_size()[2] == SizeArray[2]);
assert(Info.get_global_mem_size() == 1000);
assert(Info.get_local_mem_size() == 1000);

Expand All @@ -228,6 +223,91 @@ void test_device_info_api() {
assert(Info.get_global_mem_cache_size() == 1000);
}

void test_image_max_attrs() {
std::cout << __PRETTY_FUNCTION__ << std::endl;
syclcompat::device_info info;

int _image1d_max = 1;
int _image2d_max[2] = {2, 3};
int _image3d_max[3] = {4, 5, 6};

info.set_image1d_max(_image1d_max);
info.set_image2d_max(_image2d_max[0], _image2d_max[1]);
info.set_image3d_max(_image3d_max[0], _image3d_max[1], _image3d_max[2]);

assert(info.get_image1d_max() == _image1d_max);
assert(info.get_image2d_max()[0] == _image2d_max[0]);
assert(info.get_image2d_max()[1] == _image2d_max[1]);
assert(info.get_image3d_max()[0] == _image3d_max[0]);
assert(info.get_image3d_max()[1] == _image3d_max[1]);
assert(info.get_image3d_max()[2] == _image3d_max[2]);

DeviceExtFixt dev_ext;
auto &dev_ = dev_ext.get_dev_ext();

info.set_image1d_max(0);
info.set_image2d_max(0, 0);
info.set_image3d_max(0, 0, 0);

// SYCL guarantees at least a certain minimum value if the device has
// aspect::image
if (!dev_.has(sycl::aspect::image)) {
std::cout << " Partial skip: device does not have sycl::aspect::image."
<< std::endl;
return;
}
dev_.get_device_info(info);
// We only need to ensure the value is modified.
assert(info.get_image1d_max() > 0);
assert(info.get_image2d_max()[0] > 0);
assert(info.get_image2d_max()[1] > 0);
assert(info.get_image3d_max()[0] > 0);
assert(info.get_image3d_max()[1] > 0);
assert(info.get_image3d_max()[2] > 0);
}

void test_max_nd_range() {
std::cout << __PRETTY_FUNCTION__ << std::endl;
syclcompat::device_info info;

int size_array[3] = {1, 2, 3};
info.set_max_nd_range_size(size_array);

assert(info.get_max_nd_range_size()[0] == size_array[0]);
assert(info.get_max_nd_range_size()[1] == size_array[1]);
assert(info.get_max_nd_range_size()[2] == size_array[2]);

DeviceExtFixt dev_ext;
auto &dev = dev_ext.get_dev_ext();
dev.get_device_info(info);

int size_array_zeros[3] = {0, 0, 0};
info.set_max_nd_range_size(size_array_zeros);

#ifdef SYCL_EXT_ONEAPI_MAX_WORK_GROUP_QUERY
// According to the extension values are > 1 unless info::device_type is
// info::device_type::custom.
if (dev.get_info<sycl::info::device::device_type>() ==
sycl::info::device_type::custom) {
std::cout << " Skipping due to custom sycl::info::device_type::custom."
<< std::endl;
return;
}

info.set_max_nd_range_size(
dev.get_info<
sycl::ext::oneapi::experimental::info::device::max_work_groups<3>>());
assert(info.get_max_nd_range_size()[0] > 0);
assert(info.get_max_nd_range_size()[1] > 0);
assert(info.get_max_nd_range_size()[2] > 0);
#else
int expected = 0x7FFFFFFF;
assert(info.get_max_nd_range_size()[0] == expected);
assert(info.get_max_nd_range_size()[1] == expected);
assert(info.get_max_nd_range_size()[2] == expected);
#endif
}

int main() {
test_at_least_one_device();
test_matches_id();
Expand All @@ -242,6 +322,8 @@ int main() {
test_saved_queue();
test_reset();
test_device_info_api();
test_image_max_attrs();
test_max_nd_range();

return 0;
}

0 comments on commit b49303c

Please sign in to comment.