Developer Guide

FPGA Optimization Guide for Intel® oneAPI Toolkits

ID 767853
Date 12/16/2022
Public

A newer version of this document is available. Customers should click here to go to the newest version.

Visible to Intel only — GUID: GUID-29DA59CE-E48C-4DC1-AB90-C79DECFD6723

Document Table of Contents

Advantages and Limitations of Arbitrary Precision Data Types

Advantages

The arbitrary precision data types have the following advantages over the use of standard C/C++ data types:

  • You can achieve narrower data paths and processing elements for various operations in the circuit.
  • The data types ensure that all operations are carried out in a size guaranteed not to lose any data. However, you can still lose data if you store data in a location where the data type is too narrow in size.

Limitations

AC Data Types

The AC data types have the following limitations:

  • Multipliers are limited to generating 512-bit results.
  • Dividers for ac_int data types are limited to a maximum of 64-bit unsigned or 63-bit signed.
  • You must initialize an ac_int variable before accessing it using the bit-select operator [] or bit-slice operations slc and set_slc. Using the bit-select operator or bit-slice operations on an uninitialized ac_int variable is an undefined behavior and can give you unexpected results. Assigning each bit explicitly using the [] operator or set_slc function does not count as initializing the ac_int variable.
  • Dividers for ac_fixed data types are limited to a maximum of 64-bits (unsigned or signed).
  • Creation of ac_fixed variables larger than 32 bits are supported only with the use of the bit_fill utility function.

    For example:

    // Creating an ac_fixed with value set to 4294967298, which is larger than 2^32. // Unsupported ac_fixed<64, 64, false> v1 = ac_fixed<64, 64, false>(4294967298); // Supported // 4294967298 is 0b100000000000000000000000000000010 in binary // Express that as two 32-bit numbers and use the bit_fill utility function. const int vec_inp[2] = {0x00000001, 0x00000002}; ac_fixed<64, 64, false> bit_fill_res; bit_fill_res.bit_fill<2>(vec_inp);
  • The AC data types are not supported on the Red Hat Enterprise Linux* (RHEL) 7 operating system for emulation due to a bug in the glibc version bundled with RHEL 7.
  • You cannot template the ac_complex data type with the ap_float data type.
  • When using the bit_fill_hex() function inside a kernel, pass the input string to the kernel through a char buffer and not as a string buffer. In addition, hardware and simulation compile flows do not support using a string literal or passing the string directly to the function. The following are the supported and unsupported code patterns:

    Supported Patterns

    // Supported Pattern 1: Passing string as a char sycl::buffer to the kernel ac_int<140, false> supported_example1(queue &q) { ac_int<140, false> a; std::string hex_string{"0x177632EE7E265080BD54FF0CE7EF42C12"}; constexpr int N = 36; // size of hex_string buffer<ac_int<140, false>, 1> inp1(&a, 1); // Note: the N + 1 ensures that the null byte //terminating the char array buffer is copied buffer<char, 1> inp2(hex_string.c_str(), range<1>(N + 1)); q.submit([&](handler &h) { accessor x(inp1, h, read_write); accessor y(inp2, h, read_only); h.single_task<class D>([=] { x[0].bit_fill_hex(&y[0]); }); }); q.wait(); return a; }
    // Supported Pattern 2: Create a char array with the string literal. ac_int<140, false> supported_example2(queue &q) { ac_int<140, false> a; buffer<ac_int<140, false>, 1> inp1(&a, 1); q.submit([&](handler &h) { accessor x(inp1, h, read_write); h.single_task<class D>([=] { char str[36] = "0x177632EE7E265080BD54FF0CE7EF42C12"; x[0].bit_fill_hex(str); }); }); q.wait(); return a; }

    Unsupported Patterns

    // Unsupported Pattern 1 – Using a string Literal, will result in compilation error ac_int<140, false> unsupported_example1(queue& q) { { ac_int<140, false> a; buffer<ac_int<140, false>, 1> a_buff(&a, 1); q.submit([&](handler &h) { accessor a_acc {a_buff, h, write_only, no_init}; h.single_task<class A>([=]() { a_acc[0].bit_fill_hex("1141e98e8c51b7ac7ad387d7f8ee4f1b9"); }); }); q.wait_and_throw(); return a; } }
    // Unsupported Pattern 2 – Passing the string to the kernel in a string sycl::buffer ac_int<140, false> unsupported_example2(queue& q) { { std::string str{"1141e98e8c51b7ac7ad387d7f8ee4f1b9"}; ac_int<140, false> a; buffer<std::string, 1> str_buff(&str, 1); buffer<ac_int<140, false>, 1> a_buff(&a, 1); q.submit([&](handler &h) { accessor str_acc {str_buff, h, read_only}; accessor a_acc {a_buff, h, write_only, no_init}; h.single_task<class B>([=]() { a_acc[0].bit_fill_hex(str_acc[0].c_str()); }); }); q.wait_and_throw(); return a; } }
ap_float Data Type

The ap_float data type has the following limitations:

  • While the floating-point optimization of converting into constants is performed for float and double data types, it is not performed for the ap_float data type.
  • A limited set of math functions is supported. For details, see Math Functions Supported by ap_float Data Type.
  • Constant initialization works only with the round-towards-zero (RZERO) rounding mode.
  • For emulation, the ap_float math library is not supported on the Red Hat Enterprise Linux* (RHEL) 7 operating system.
  • When computing A^B using ap_float's ihc_pown function, if B is an unsigned type T of size N bits and is equal to the maximum unsigned value, redefine B to be of size N+1 bits. Otherwise, results will be incorrect. For example:
    // Sample Code: ap_float<8, 7> a = 2; ac_int<4, false> b = 15; // max value that this ac_int can hold = ihc_pown(a , b); // !!! Will produce incorrect result
    // Workaround: ap_float<8, 7> a = 2; ac_int<5, false> b = 15; // Workaround = ihc_pown(a , b); // Will produce correct result