DPC++ Runtime
Runtime libraries for oneAPI DPC++
spirv_types.hpp
Go to the documentation of this file.
1 //===------------ spirv_types.hpp --- SPIRV types -------------------------===//
2 //
3 // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
4 // See https://llvm.org/LICENSE.txt for license information.
5 // SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
6 //
7 //===----------------------------------------------------------------------===//
8 
9 #pragma once
10 
11 #include "sycl/half_type.hpp"
12 #include <sycl/detail/defines.hpp>
14 #include <sycl/half_type.hpp>
15 
16 #include <complex>
17 #include <cstddef>
18 #include <cstdint>
19 
20 // TODO: include the header file with SPIR-V declarations from SPIRV-Headers
21 // project.
22 
23 // Declarations of enums below is aligned with corresponding declarations in
24 // SPIRV-Headers repo with a few exceptions:
25 // - base types changed from uint to uint32_t
26 // - spv namespace renamed to __spv
27 namespace __spv {
28 
29 struct Scope {
30 
31  enum Flag : uint32_t {
33  Device = 1,
34  Workgroup = 2,
35  Subgroup = 3,
37  };
38 
39  constexpr Scope(Flag flag) : flag_value(flag) {}
40 
41  constexpr operator uint32_t() const { return flag_value; }
42 
44 };
45 
46 struct StorageClass {
47  enum Flag : uint32_t {
49  Input = 1,
50  Uniform = 2,
51  Output = 3,
52  Workgroup = 4,
54  Private = 6,
55  Function = 7,
56  Generic = 8,
59  Image = 11,
65  RayPayloadKHR = 5338,
66  RayPayloadNV = 5338,
78  HostOnlyINTEL = 5937,
79  Max = 0x7fffffff,
80  };
81  constexpr StorageClass(Flag flag) : flag_value(flag) {}
82  constexpr operator uint32_t() const { return flag_value; }
84 };
85 
87 
88  enum Flag : uint32_t {
89  None = 0x0,
90  Acquire = 0x2,
91  Release = 0x4,
94  UniformMemory = 0x40,
96  WorkgroupMemory = 0x100,
99  ImageMemory = 0x800,
100  };
101 
102  constexpr MemorySemanticsMask(Flag flag) : flag_value(flag) {}
103 
104  constexpr operator uint32_t() const { return flag_value; }
105 
107 };
108 
109 enum class GroupOperation : uint32_t {
110  Reduce = 0,
111  InclusiveScan = 1,
112  ExclusiveScan = 2
113 };
114 
115 #if (SYCL_EXT_ONEAPI_MATRIX_VERSION > 1)
116 enum class MatrixLayout : uint32_t {
117  RowMajor = 0,
118  ColumnMajor = 1,
119  Packed = 2,
120  Dynamic = 3
121 };
122 #else
123 enum class MatrixLayout : uint32_t {
124  RowMajor = 0,
125  ColumnMajor = 1,
126  PackedA = 2,
127  PackedB = 3,
128  Unused = 4
129 };
130 #endif
131 
132 enum class MatrixUse : uint32_t { MatrixA = 0, MatrixB = 1, Accumulator = 2 };
133 
135  complex_float() = default;
136  complex_float(std::complex<float> x) : real(x.real()), imag(x.imag()) {}
137  operator std::complex<float>() { return {real, imag}; }
138  float real, imag;
139 };
140 
142  complex_double() = default;
143  complex_double(std::complex<double> x) : real(x.real()), imag(x.imag()) {}
144  operator std::complex<double>() { return {real, imag}; }
145  double real, imag;
146 };
147 
148 struct complex_half {
149  complex_half() = default;
150  complex_half(std::complex<sycl::half> x) : real(x.real()), imag(x.imag()) {}
151  operator std::complex<sycl::half>() { return {real, imag}; }
153 };
154 
155 #if (SYCL_EXT_ONEAPI_MATRIX_VERSION > 1)
156 template <typename T, std::size_t R, std::size_t C, MatrixLayout L,
157  Scope::Flag S = Scope::Flag::Subgroup,
160 #else
161 template <typename T, std::size_t R, std::size_t C, MatrixLayout L,
162  Scope::Flag S = Scope::Flag::Subgroup>
164 #endif // SYCL_EXT_ONEAPI_MATRIX_VERSION
165 
166 } // namespace __spv
167 
168 #ifdef __SYCL_DEVICE_ONLY__
169 // OpenCL pipe types
170 template <typename dataT>
171 using __ocl_RPipeTy = __attribute__((pipe("read_only"))) const dataT;
172 template <typename dataT>
173 using __ocl_WPipeTy = __attribute__((pipe("write_only"))) const dataT;
174 
175 // OpenCL vector types
176 template <typename dataT, int dims>
177 using __ocl_vec_t = dataT __attribute__((ext_vector_type(dims)));
178 
179 // Struct representing layout of pipe storage
180 // TODO: rename to __spirv_ConstantPipeStorage
181 struct ConstantPipeStorage {
182  int32_t _PacketSize;
183  int32_t _PacketAlignment;
184  int32_t _Capacity;
185 };
186 
187 namespace sycl {
189 namespace detail {
190 // Arbitrary precision integer type
191 template <int Bits> using ap_int = _BitInt(Bits);
192 } // namespace detail
193 } // __SYCL_INLINE_VER_NAMESPACE(_V1)
194 } // namespace sycl
195 #endif // __SYCL_DEVICE_ONLY__
196 
197 // This class does not have definition, it is only predeclared here.
198 // The pointers to this class objects can be passed to or returned from
199 // SPIRV built-in functions.
200 // Only in such cases the class is recognized as SPIRV type __ocl_event_t.
201 #ifndef __SYCL_DEVICE_ONLY__
202 typedef void* __ocl_event_t;
203 typedef void* __ocl_sampler_t;
204 // Adding only the datatypes that can be currently used in SYCL,
205 // as per SYCL spec 1.2.1
206 #define __SYCL_SPV_IMAGE_TYPE(NAME) typedef void *__ocl_##NAME##_t
207 
208 #define __SYCL_SPV_SAMPLED_AND_IMAGE_TYPE(NAME) \
209  __SYCL_SPV_IMAGE_TYPE(NAME); \
210  typedef void *__ocl_sampled_##NAME##_t
211 
215 __SYCL_SPV_IMAGE_TYPE(image1d_wo);
216 __SYCL_SPV_IMAGE_TYPE(image2d_wo);
217 __SYCL_SPV_IMAGE_TYPE(image3d_wo);
218 __SYCL_SPV_SAMPLED_AND_IMAGE_TYPE(image1d_array_ro);
219 __SYCL_SPV_SAMPLED_AND_IMAGE_TYPE(image2d_array_ro);
220 __SYCL_SPV_IMAGE_TYPE(image1d_array_wo);
221 __SYCL_SPV_IMAGE_TYPE(image2d_array_wo);
222 
223 #undef __SYCL_SPV_IMAGE_TYPE
224 #undef __SYCL_SPV_SAMPLED_AND_IMAGE_TYPE
225 #endif
__spv::StorageClass::Function
@ Function
Definition: spirv_types.hpp:55
__spv::StorageClass::AtomicCounter
@ AtomicCounter
Definition: spirv_types.hpp:58
__spv::complex_half::real
sycl::half real
Definition: spirv_types.hpp:152
__spv::MemorySemanticsMask::SubgroupMemory
@ SubgroupMemory
Definition: spirv_types.hpp:95
__spv::complex_float::complex_float
complex_float(std::complex< float > x)
Definition: spirv_types.hpp:136
__spv::Scope::flag_value
Flag flag_value
Definition: spirv_types.hpp:43
__spv::complex_double::imag
double imag
Definition: spirv_types.hpp:145
__spv::Scope::Workgroup
@ Workgroup
Definition: spirv_types.hpp:34
__spv::complex_float::complex_float
complex_float()=default
__spv::MatrixLayout::Dynamic
@ Dynamic
__spv::complex_double::complex_double
complex_double(std::complex< double > x)
Definition: spirv_types.hpp:143
__spv::StorageClass::IncomingRayPayloadKHR
@ IncomingRayPayloadKHR
Definition: spirv_types.hpp:69
sycl::_V1::pipe
ext::intel::pipe< name, dataT, min_capacity > pipe
Definition: pipes.hpp:16
__spv::StorageClass::CodeSectionINTEL
@ CodeSectionINTEL
Definition: spirv_types.hpp:75
__spv::MatrixLayout::RowMajor
@ RowMajor
__spv::StorageClass::IncomingCallableDataKHR
@ IncomingCallableDataKHR
Definition: spirv_types.hpp:63
__spv::MemorySemanticsMask::AtomicCounterMemory
@ AtomicCounterMemory
Definition: spirv_types.hpp:98
__spv
Definition: spirv_types.hpp:27
__SYCL_SPV_SAMPLED_AND_IMAGE_TYPE
#define __SYCL_SPV_SAMPLED_AND_IMAGE_TYPE(NAME)
Definition: spirv_types.hpp:208
__SYCL_INLINE_VER_NAMESPACE
#define __SYCL_INLINE_VER_NAMESPACE(X)
Definition: defines_elementary.hpp:11
__spv::StorageClass::ShaderRecordBufferNV
@ ShaderRecordBufferNV
Definition: spirv_types.hpp:72
__spv::StorageClass::PhysicalStorageBufferEXT
@ PhysicalStorageBufferEXT
Definition: spirv_types.hpp:74
__spv::StorageClass::CapabilityUSMStorageClassesINTEL
@ CapabilityUSMStorageClassesINTEL
Definition: spirv_types.hpp:76
__spv::StorageClass::CallableDataNV
@ CallableDataNV
Definition: spirv_types.hpp:62
__spv::MemorySemanticsMask::MemorySemanticsMask
constexpr MemorySemanticsMask(Flag flag)
Definition: spirv_types.hpp:102
__spv::StorageClass::HitAttributeKHR
@ HitAttributeKHR
Definition: spirv_types.hpp:67
__spv::complex_half::imag
sycl::half imag
Definition: spirv_types.hpp:152
__spv::StorageClass
Definition: spirv_types.hpp:46
__spv::MemorySemanticsMask::Acquire
@ Acquire
Definition: spirv_types.hpp:90
__spv::StorageClass::HostOnlyINTEL
@ HostOnlyINTEL
Definition: spirv_types.hpp:78
__spv::__spirv_JointMatrixINTEL
Definition: spirv_types.hpp:159
sycl
---— Error handling, matching OpenCL plugin semantics.
Definition: access.hpp:14
__spv::MemorySemanticsMask::AcquireRelease
@ AcquireRelease
Definition: spirv_types.hpp:92
__spv::StorageClass::Image
@ Image
Definition: spirv_types.hpp:59
__spv::MemorySemanticsMask::CrossWorkgroupMemory
@ CrossWorkgroupMemory
Definition: spirv_types.hpp:97
__ocl_event_t
void * __ocl_event_t
Definition: spirv_types.hpp:202
__spv::MatrixUse
MatrixUse
Definition: spirv_types.hpp:132
__spv::complex_float::imag
float imag
Definition: spirv_types.hpp:138
__spv::StorageClass::CallableDataKHR
@ CallableDataKHR
Definition: spirv_types.hpp:61
__spv::complex_half::complex_half
complex_half()=default
__spv::StorageClass::HitAttributeNV
@ HitAttributeNV
Definition: spirv_types.hpp:68
__spv::Scope::Flag
Flag
Definition: spirv_types.hpp:31
__spv::StorageClass::IncomingRayPayloadNV
@ IncomingRayPayloadNV
Definition: spirv_types.hpp:70
__spv::StorageClass::Flag
Flag
Definition: spirv_types.hpp:47
__spv::MemorySemanticsMask::flag_value
Flag flag_value
Definition: spirv_types.hpp:106
__spv::StorageClass::Generic
@ Generic
Definition: spirv_types.hpp:56
__spv::GroupOperation::ExclusiveScan
@ ExclusiveScan
__spv::complex_double::complex_double
complex_double()=default
__spv::MemorySemanticsMask
Definition: spirv_types.hpp:86
defines_elementary.hpp
__spv::complex_double::real
double real
Definition: spirv_types.hpp:145
__spv::StorageClass::IncomingCallableDataNV
@ IncomingCallableDataNV
Definition: spirv_types.hpp:64
__spv::Scope::Subgroup
@ Subgroup
Definition: spirv_types.hpp:35
__spv::StorageClass::PhysicalStorageBuffer
@ PhysicalStorageBuffer
Definition: spirv_types.hpp:73
__spv::StorageClass::Uniform
@ Uniform
Definition: spirv_types.hpp:50
__spv::complex_float::real
float real
Definition: spirv_types.hpp:138
__spv::StorageClass::Private
@ Private
Definition: spirv_types.hpp:54
__spv::GroupOperation::InclusiveScan
@ InclusiveScan
__spv::MemorySemanticsMask::ImageMemory
@ ImageMemory
Definition: spirv_types.hpp:99
defines.hpp
__spv::StorageClass::PushConstant
@ PushConstant
Definition: spirv_types.hpp:57
__spv::MatrixUse::MatrixA
@ MatrixA
__spv::Scope::Invocation
@ Invocation
Definition: spirv_types.hpp:36
sycl::_V1::half
sycl::detail::half_impl::half half
Definition: aliases.hpp:103
__spv::Scope::Scope
constexpr Scope(Flag flag)
Definition: spirv_types.hpp:39
__spv::StorageClass::Max
@ Max
Definition: spirv_types.hpp:79
__spv::MatrixUse::MatrixB
@ MatrixB
__spv::complex_float
Definition: spirv_types.hpp:134
__spv::MemorySemanticsMask::SequentiallyConsistent
@ SequentiallyConsistent
Definition: spirv_types.hpp:93
sycl::_V1::write_only
constexpr mode_tag_t< access_mode::write > write_only
Definition: access.hpp:75
__spv::MemorySemanticsMask::None
@ None
Definition: spirv_types.hpp:89
__SYCL_SPV_IMAGE_TYPE
#define __SYCL_SPV_IMAGE_TYPE(NAME)
Definition: spirv_types.hpp:206
__spv::StorageClass::CrossWorkgroup
@ CrossWorkgroup
Definition: spirv_types.hpp:53
__spv::MemorySemanticsMask::WorkgroupMemory
@ WorkgroupMemory
Definition: spirv_types.hpp:96
__spv::StorageClass::Workgroup
@ Workgroup
Definition: spirv_types.hpp:52
__spv::MatrixLayout::ColumnMajor
@ ColumnMajor
__spv::StorageClass::ShaderRecordBufferKHR
@ ShaderRecordBufferKHR
Definition: spirv_types.hpp:71
__spv::MatrixUse::Accumulator
@ Accumulator
__spv::StorageClass::UniformConstant
@ UniformConstant
Definition: spirv_types.hpp:48
__spv::MemorySemanticsMask::Release
@ Release
Definition: spirv_types.hpp:91
__spv::Scope::CrossDevice
@ CrossDevice
Definition: spirv_types.hpp:32
__spv::StorageClass::flag_value
Flag flag_value
Definition: spirv_types.hpp:83
__spv::MemorySemanticsMask::Flag
Flag
Definition: spirv_types.hpp:88
__spv::StorageClass::RayPayloadKHR
@ RayPayloadKHR
Definition: spirv_types.hpp:65
__spv::StorageClass::Input
@ Input
Definition: spirv_types.hpp:49
__spv::MatrixLayout
MatrixLayout
Definition: spirv_types.hpp:116
__spv::StorageClass::DeviceOnlyINTEL
@ DeviceOnlyINTEL
Definition: spirv_types.hpp:77
half_type.hpp
__spv::GroupOperation
GroupOperation
Definition: spirv_types.hpp:109
__spv::StorageClass::Output
@ Output
Definition: spirv_types.hpp:51
__spv::Scope
Definition: spirv_types.hpp:29
__spv::StorageClass::RayPayloadNV
@ RayPayloadNV
Definition: spirv_types.hpp:66
__spv::MatrixLayout::Packed
@ Packed
__spv::complex_double
Definition: spirv_types.hpp:141
__spv::complex_half::complex_half
complex_half(std::complex< sycl::half > x)
Definition: spirv_types.hpp:150
__spv::StorageClass::StorageBuffer
@ StorageBuffer
Definition: spirv_types.hpp:60
__spv::GroupOperation::Reduce
@ Reduce
sycl::_V1::ext::oneapi::experimental::__attribute__
__attribute__((always_inline)) auto invoke_simd(sycl
The invoke_simd free function invokes a SIMD function using all work-items in a sub_group.
Definition: invoke_simd.hpp:357
__spv::StorageClass::StorageClass
constexpr StorageClass(Flag flag)
Definition: spirv_types.hpp:81
__spv::Scope::Device
@ Device
Definition: spirv_types.hpp:33
__spv::complex_half
Definition: spirv_types.hpp:148
__ocl_sampler_t
void * __ocl_sampler_t
Definition: spirv_types.hpp:203
__spv::MemorySemanticsMask::UniformMemory
@ UniformMemory
Definition: spirv_types.hpp:94