Design support for reduced precision types in dpctl

[diptorupd]

We should support reduced precision types that are supported in Sycl.

[PokhodenkoSA]

• half (symbol "e") is supported in module struct starting from Python 3.6 (https://docs.python.org/3/library/struct.html#format-characters)
• half is not supported in ctypes. (https://docs.python.org/3/library/ctypes.html?highlight=ctypes#fundamental-data-types) Error:

class c_half(ctypes._SimpleCData):
  _type_ = "e"

Traceback (most recent call last):
  File "<stdin>", line 1, in <module>
AttributeError: class must define a '_type_' attribute which must be
a single character string containing one of 'cbBhHiIlLdfuzZqQPXOv?g'.

• half/float16 is supported in NumPy (https://numpy.org/doc/stable/reference/arrays.scalars.html#numpy.float16)
• float16 is supported in PyArrow (https://arrow.apache.org/docs/python/generated/pyarrow.float16.html#pyarrow.float16)
• np.float16 is not supported in Numba (https://numba.pydata.org/numba-doc/dev/reference/numpysupported.html?highlight=half#scalar-types)

Links:

• Half-precision floating-point format (Wiki)
• bfloat16 floating-point format
• BFLOAT16 – Hardware Numerics Definition
• OpenVINO Bfloat16 Inference
• Brain floating-point format (bfloat16) (Wikichip)
• https://stackoverflow.com/questions/49995594/half-precision-floating-point-arithmetic-on-intel-chips
• https://software.intel.com/content/www/us/en/develop/articles/performance-benefits-of-half-precision-floats.html
• https://software.intel.com/content/www/us/en/develop/blogs/intel-half-precision-floating-point-format-conversion-instructions.html
• https://software.intel.com/content/www/us/en/develop/articles/intel-deep-learning-boost-new-instruction-bfloat16.html
• https://scicomp.stackexchange.com/questions/35187/is-half-precision-supported-by-modern-architecture
• https://stackoverflow.com/questions/5766882/why-is-there-no-2-byte-float-and-does-an-implementation-already-exist
• From DAAL team: https://github.com/rlnx/reduced-precision-for-dba
• https://cloud.google.com/tpu/docs/bfloat16#the_bfloat16_floating-point_format

[oleksandr-pavlyk]

The scope of the support matters. The main purpose of supporting in dpctl is to be able to submit kernels that take reduced precision arguments.

In addition to standard C/C++ types, SYCL supports std::byte and sycl::half (assuming device supports it). See tables 4.105 and 5.1 in the SYCL 2020 provisional standard. bfloa16 is not presently supported. Reference to an extension proposal would be good to record in this issue.

Anyhow, for dpctl to support any additional type, we need Python class representing the type, say dpctl.types.c_half and dpctl.types.c_bfloat16, whose purpose in life is to hold the value, and designate its type. It should have a constructor to convert from other floating point types, i.e. Python floats, ctypes.c_float, ctypes.c_double, etc., and allow to get the reference to the value.

DPCTL's SyclQueue.submit will then recognize this class, and call KernelSetArg with the proper enum value.

[oleksandr-pavlyk]

Here is example of using ctypes to convert a single ctype.c_float value input into a ctypes.c_short encoding the bits of the corresponding float16 value. https://gamedev.stackexchange.com/questions/28023/python-float-32bit-to-half-float-16bit

I wonder is this could be a start of implementing our own float16 type object.

[diptorupd]

@oleksandr-pavlyk Is the issue still relevant? If not I will suggest it to be closed.