diff --git a/docs/backend/oneapi.rst b/docs/backend/oneapi.rst index 585bfc27cb..58904894f3 100644 --- a/docs/backend/oneapi.rst +++ b/docs/backend/oneapi.rst @@ -1,24 +1,28 @@ -====== -oneAPI -====== +=========================== +Altera HLS IP Gen / oneAPI +=========================== -The **oneAPI** backend of hls4ml is designed for deploying NNs on Intel/Altera FPGAs. It will eventually -replace the **Quartus** backend, which targeted Intel HLS. (Quartus continues to be used with IP produced by the -**oneAPI** backend.) This section discusses details of the **oneAPI** backend. +The Altera HLS/oneAPI backend of hls4ml is designed for deploying neural networks on Intel/Altera FPGAs. +It is based on the SYCL FPGA flow that was previously provided through **Intel oneAPI** and is now continued +by the **Altera HLS IP Gen** toolchain. The functionality and terminology used by hls4ml remain largely unchanged. -The **oneAPI** code uses SYCL kernels to implement the logic that is deployed on FPGAs. It naturally leads to the +The **Altera HLS/oneAPI** backend is intended to supersede the **Quartus** backend, which targets the discontinued +Intel HLS Compiler. Quartus Prime continues to be used to integrate the IP generated by the Altera HLS IP Gen flow +into FPGA designs. + +The **Altera HLS/oneAPI** code uses SYCL kernels to implement the logic that is deployed on FPGAs. It naturally leads to the accelerator style of programming. In the SYCL HLS (IP Component) flow, which is currently the only flow supported, the kernel becomes the IP, and the "host code" becomes the testbench. An accelerator flow, with easier deployment on PCIe accelerator boards, is planned to be added in the future. The produced work areas use cmake to build the projects in a style based -`oneAPI-samples `_. +`Altera HLS IP Gen hls-samples `_ (or deprecated `oneAPI-samples `_). The standard ``fpga_emu``, ``report``, ``fpga_sim``, and ``fpga`` make targets are supported. Additionally, ``make lib`` produces the library used for calling the ``predict`` function from hls4ml. The ``compile`` and ``build`` commands in hls4ml interact with the cmake system, so one does not need to manually use the build system, but it there if desired. -The **oneAPI** backend, like the **Quartus** backend, only implements the ``Resource`` strategy for the layers. There +The **Altera HLS/oneAPI** backend, like the **Quartus** backend, only implements the ``Resource`` strategy for the layers. There is no ``Latency`` implementation of any of the layers. Note: currently tracing and external weights (i.e. setting BramFactor) are not supported. @@ -27,7 +31,7 @@ io_parallel and io_stream ========================= As mentioned in the :ref:`I/O Types` section, ``io_parallel`` is for small models, while ``io_stream`` is for -larger models. In ``oneAPI``, there is an additional difference: ``io_stream`` implements each layer on its +larger models. In ``Altera HLS/oneAPI``, there is an additional difference: ``io_stream`` implements each layer on its own ``task_sequence``. Thus, the layers run in parallel, with pipes connecting the inputs and outputs. This is similar in style to the `dataflow` implementation on Vitis HLS, but more explicit. It is also a change relative to the Intel HLS-based ``Quartus`` backend. On the other hand, ``io_parallel`` always uses a single task, diff --git a/docs/backend/quartus.rst b/docs/backend/quartus.rst index 8cde5f97b2..bd509e5e1d 100644 --- a/docs/backend/quartus.rst +++ b/docs/backend/quartus.rst @@ -3,10 +3,10 @@ Quartus ======= .. warning:: - The **Quartus** backend is deprecated and will be removed in a future version. Users should migrate to the **oneAPI** backend. + The **Quartus** backend is deprecated and will be removed in a future version. Users should migrate to the **Altera HLS/oneAPI** backend. -The **Quartus** backend of hls4ml is designed for deploying NNs on Intel/Altera FPGAs. It uses the discontinued Intel HLS compiler. The **oneAPI** backend should be preferred for new projects. -The **oneAPI** backend contains the migrated the HLS code from this backend, with significantly better io_stream support, though the **oneAPI** backend does not yet support profiling, tracing, +The **Quartus** backend of hls4ml is designed for deploying NNs on Intel/Altera FPGAs. It uses the discontinued Intel HLS compiler. The **Altera HLS/oneAPI** backend should be preferred for new projects. +The **Altera HLS/oneAPI** backend contains the migrated the HLS code from this backend, with significantly better io_stream support, though the **Altera HLS/oneAPI** backend does not yet support profiling, tracing, or the BramFactor option supported by the **Quartus** backend. Nevertheless, little or no further development is expected for this backend. The **Quartus** backend only implements the ``Resource`` strategy for the layers. There is no ``Latency`` implementation of any of the layers. diff --git a/docs/intro/status.rst b/docs/intro/status.rst index d3942691ca..16e6f4b701 100644 --- a/docs/intro/status.rst +++ b/docs/intro/status.rst @@ -47,6 +47,7 @@ HLS backends: * Intel HLS * Vitis HLS * Catapult HLS +* Altera HLS (experimental) * oneAPI (experimental) * XLS (experimental) @@ -54,37 +55,39 @@ A summary of the on-going status of the ``hls4ml`` tool is in the table below. .. table:: hls4ml Supported Features -+-----------------------+-----+-----+--------------+--------+--------+-----+ -| Frontend/Backend | MLP | CNN | RNN/LSTM/GRU | GarNet | Einsum | MHA | -+=======================+=====+=====+==============+========+========+=====+ -| Keras v2 | ✅ | ✅ | ✅ | ✅ | ❌ | ❌ | -+-----------------------+-----+-----+--------------+--------+--------+-----+ -| QKeras | ✅ | ✅ | ✅ | ✅ | N/A | N/A | -+-----------------------+-----+-----+--------------+--------+--------+-----+ -| QKeras-v3 | ✅ | ✅ | ✅ | ✅ | ✅ | N/A | -+-----------------------+-----+-----+--------------+--------+--------+-----+ -| HGQ | ✅ | ✅ | N/A | N/A | N/A | N/A | -+-----------------------+-----+-----+--------------+--------+--------+-----+ -| Keras v3 | ✅ | ✅ | ✅ | N/A | ✅ | ❌ | -+-----------------------+-----+-----+--------------+--------+--------+-----+ -| HGQ2 | ✅ | ✅ | N/A | N/A | ✅ | ✅ | -+-----------------------+-----+-----+--------------+--------+--------+-----+ -| Torch | ✅ | ✅ | ✅ | ❌ | ✅ | ❌ | -+-----------------------+-----+-----+--------------+--------+--------+-----+ -| ONNX | ✅ | ✅ | ❌ | ❌ | ❌ | ❌ | -+-----------------------+-----+-----+--------------+--------+--------+-----+ -| QONNX | ✅ | ✅ | ❌ | N/A | N/A | N/A | -+-----------------------+-----+-----+--------------+--------+--------+-----+ -| Vivado/Vitis HLS | ✅ | ✅ | ✅ | ❌ | ✅ | ✅ | -+-----------------------+-----+-----+--------------+--------+--------+-----+ -| Intel HLS | ✅ | ✅ | ✅ | ❌ | ❌ | ❌ | -+-----------------------+-----+-----+--------------+--------+--------+-----+ -| Catapult HLS | ✅ | ✅ | ✅ | ❌ | ❌ | ❌ | -+-----------------------+-----+-----+--------------+--------+--------+-----+ -| oneAPI (experimental) | ✅ | ✅ | ✅ | ❌ | ✅ | ❌ | -+-----------------------+-----+-----+--------------+--------+--------+-----+ -| XLS (experimental) | ✅ | ✅ | ❌ | ❌ | ❌ | ❌ | -+-----------------------+-----+-----+--------------+--------+--------+-----+ ++---------------------------+-----+-----+--------------+--------+--------+-----+ +| Frontend/Backend | MLP | CNN | RNN/LSTM/GRU | GarNet | Einsum | MHA | ++===========================+=====+=====+==============+========+========+=====+ +| Keras v2 | ✅ | ✅ | ✅ | ✅ | ❌ | ❌ | ++---------------------------+-----+-----+--------------+--------+--------+-----+ +| QKeras | ✅ | ✅ | ✅ | ✅ | N/A | N/A | ++---------------------------+-----+-----+--------------+--------+--------+-----+ +| QKeras-v3 | ✅ | ✅ | ✅ | ✅ | ✅ | N/A | ++---------------------------+-----+-----+--------------+--------+--------+-----+ +| HGQ | ✅ | ✅ | N/A | N/A | N/A | N/A | ++---------------------------+-----+-----+--------------+--------+--------+-----+ +| Keras v3 | ✅ | ✅ | ✅ | N/A | ✅ | ❌ | ++---------------------------+-----+-----+--------------+--------+--------+-----+ +| HGQ2 | ✅ | ✅ | N/A | N/A | ✅ | ✅ | ++---------------------------+-----+-----+--------------+--------+--------+-----+ +| Torch | ✅ | ✅ | ✅ | ❌ | ✅ | ❌ | ++---------------------------+-----+-----+--------------+--------+--------+-----+ +| ONNX | ✅ | ✅ | ❌ | ❌ | ❌ | ❌ | ++---------------------------+-----+-----+--------------+--------+--------+-----+ +| QONNX | ✅ | ✅ | ❌ | N/A | N/A | N/A | ++---------------------------+-----+-----+--------------+--------+--------+-----+ +| Vivado/Vitis HLS | ✅ | ✅ | ✅ | ❌ | ✅ | ✅ | ++---------------------------+-----+-----+--------------+--------+--------+-----+ +| Intel HLS | ✅ | ✅ | ✅ | ❌ | ❌ | ❌ | ++---------------------------+-----+-----+--------------+--------+--------+-----+ +| Catapult HLS | ✅ | ✅ | ✅ | ❌ | ❌ | ❌ | ++---------------------------+-----+-----+--------------+--------+--------+-----+ +| Altera HLS (experimental) | ✅ | ✅ | ✅ | ❌ | ✅ | ❌ | ++---------------------------+-----+-----+--------------+--------+--------+-----+ +| oneAPI (experimental) | ✅ | ✅ | ✅ | ❌ | ✅ | ❌ | ++---------------------------+-----+-----+--------------+--------+--------+-----+ +| XLS (experimental) | ✅ | ✅ | ❌ | ❌ | ❌ | ❌ | ++---------------------------+-----+-----+--------------+--------+--------+-----+ Other feature notes: @@ -94,6 +97,7 @@ Other feature notes: - Intel HLS versions 20.1 to 21.4, versions > 21.4 have not been tested. - Vitis HLS versions 2022.2 to 2024.1. Versions > 2024.1 are less tested. - Catapult HLS versions 2024.1_1 to 2024.2 + - Altera HLS IP Gen (Beta) 2026.1 - oneAPI versions 2024.1 to 2025.0. Any future versions are known to not work. * ``hls4ml`` supports Linux [*]_ and requires python >=3.10. hls4ml does not require a specific Linux distribution version and we recommend following the requirements of the HLS tool you are using. diff --git a/hls4ml/backends/oneapi/oneapi_backend.py b/hls4ml/backends/oneapi/oneapi_backend.py index 0c11c16d09..cb317c2e61 100644 --- a/hls4ml/backends/oneapi/oneapi_backend.py +++ b/hls4ml/backends/oneapi/oneapi_backend.py @@ -208,9 +208,12 @@ def build(self, model, build_type='fpga_emu', run=False): builddir = outdir / 'build' builddir.mkdir(exist_ok=True) try: - subprocess.run('which icpx', shell=True, cwd=builddir, check=True) + subprocess.run('which ahls', shell=True, cwd=builddir, check=True) except subprocess.CalledProcessError: - raise RuntimeError('Could not find icpx. Please configure oneAPI appropriately') + try: + subprocess.run('which icpx', shell=True, cwd=builddir, check=True) + except subprocess.CalledProcessError: + raise RuntimeError('Could not find Altera HLS or oneAPI. Please configure your environment appropriately.') subprocess.run('cmake ..', shell=True, cwd=builddir, check=True) subprocess.run(f'make {build_type}', shell=True, cwd=builddir, check=True) diff --git a/hls4ml/backends/oneapi/oneapi_types.py b/hls4ml/backends/oneapi/oneapi_types.py index 0ecff79034..765a258c44 100644 --- a/hls4ml/backends/oneapi/oneapi_types.py +++ b/hls4ml/backends/oneapi/oneapi_types.py @@ -181,9 +181,15 @@ def definition_cpp(self, name_suffix='', as_reference=False): def declare_cpp(self, pipe_min_size=0, indent=''): lines = indent + f'class {self.pipe_id};\n' + # This will look very ugly but it is the only way to ensure synch with CMake compiler decision lines += indent + ( + '#ifdef AHLS\n' + f'using {self.pipe_name} = sycl::ext::altera::experimental::pipe<{self.pipe_id}, ' + f'{self.type.name}, {pipe_min_size}, PipeProps>;\n' + '#else\n' f'using {self.pipe_name} = sycl::ext::intel::experimental::pipe<{self.pipe_id}, ' - + f'{self.type.name}, {pipe_min_size}, PipeProps>;\n' + f'{self.type.name}, {pipe_min_size}, PipeProps>;\n' + '#endif\n' ) return lines @@ -204,8 +210,13 @@ def definition_cpp(self, name_suffix='', as_reference=True): def declare_cpp(self, indent=''): lines = indent + f'class {self.pipe_id};\n' lines += indent + ( + '#ifdef AHLS\n' + f'using {self.pipe_name} = sycl::ext::altera::experimental::pipe<{self.pipe_id}, ' + f'{self.type.name}, {self.pragma[-1]}>;\n' + '#else\n' f'using {self.pipe_name} = sycl::ext::intel::experimental::pipe<{self.pipe_id}, ' - + f'{self.type.name}, {self.pragma[-1]}>;\n' + f'{self.type.name}, {self.pragma[-1]}>;\n' + '#endif\n' ) return lines diff --git a/hls4ml/templates/oneapi/CMakeLists.txt b/hls4ml/templates/oneapi/CMakeLists.txt index 5bce2aaf84..b79790e479 100644 --- a/hls4ml/templates/oneapi/CMakeLists.txt +++ b/hls4ml/templates/oneapi/CMakeLists.txt @@ -1,7 +1,12 @@ -# Direct CMake to use icpx rather than the default C++ compiler/linker on Linux +# Direct CMake to use ahls/icpx rather than the default C++ compiler/linker on Linux # and icx-cl on Windows if(UNIX) - set(CMAKE_CXX_COMPILER icpx) + find_program(CXX_COMPILER NAMES ahls icpx) + if(NOT CXX_COMPILER) + message(FATAL_ERROR "Neither icpx nor ahls was found in PATH.") + endif() + set(CMAKE_CXX_COMPILER "${CXX_COMPILER}") + message(STATUS "Using C++ compiler: ${CXX_COMPILER}") else() # Windows include (CMakeForceCompiler) CMAKE_FORCE_CXX_COMPILER (icx-cl IntelDPCPP) @@ -12,6 +17,10 @@ cmake_minimum_required (VERSION 3.7.2) project(myproject CXX) +if(CXX_COMPILER MATCHES "ahls$") + add_compile_definitions(AHLS) +endif() + set(CMAKE_ARCHIVE_OUTPUT_DIRECTORY ${CMAKE_BINARY_DIR}) set(CMAKE_LIBRARY_OUTPUT_DIRECTORY ${CMAKE_BINARY_DIR}) set(CMAKE_RUNTIME_OUTPUT_DIRECTORY ${CMAKE_BINARY_DIR}) diff --git a/hls4ml/templates/oneapi/firmware/defines.h b/hls4ml/templates/oneapi/firmware/defines.h index 05de507dcd..26f408f685 100644 --- a/hls4ml/templates/oneapi/firmware/defines.h +++ b/hls4ml/templates/oneapi/firmware/defines.h @@ -1,9 +1,15 @@ #ifndef DEFINES_H_ #define DEFINES_H_ +#ifdef AHLS +#include +#include +#include +#else #include #include #include +#endif #include // Include nnet::array - a custom array-like struct, mainly used with io_stream diff --git a/hls4ml/templates/oneapi/firmware/myproject.cpp b/hls4ml/templates/oneapi/firmware/myproject.cpp index 06e7d3fe37..d345484793 100644 --- a/hls4ml/templates/oneapi/firmware/myproject.cpp +++ b/hls4ml/templates/oneapi/firmware/myproject.cpp @@ -1,13 +1,21 @@ #include "myproject.h" #include "parameters.h" +#ifdef AHLS +#include +#else #include +#endif // hls-fpga-machine-learning insert weights // The inter-task pipes need to be declared in the global scope // hls-fpga-machine-learning insert inter-task pipes +#ifdef AHLS +using sycl::ext::altera::experimental::task_sequence; +#else using sycl::ext::intel::experimental::task_sequence; +#endif void MyProject::operator()() const { // **************************************** diff --git a/hls4ml/templates/oneapi/firmware/myproject.h b/hls4ml/templates/oneapi/firmware/myproject.h index 082ae5dc8c..fc1ae92665 100644 --- a/hls4ml/templates/oneapi/firmware/myproject.h +++ b/hls4ml/templates/oneapi/firmware/myproject.h @@ -6,7 +6,14 @@ // This file defines the interface to the kernel // currently this is fixed -using PipeProps = decltype(sycl::ext::oneapi::experimental::properties(sycl::ext::intel::experimental::ready_latency<0>)); + +using PipeProps = decltype(sycl::ext::oneapi::experimental::properties( +#ifdef AHLS + sycl::ext::altera::experimental::ready_latency<0> +#else + sycl::ext::intel::experimental::ready_latency<0> +#endif + )); // Need to declare the input and output pipes @@ -19,8 +26,13 @@ struct MyProject { // kernel property method to config invocation interface auto get(sycl::ext::oneapi::experimental::properties_tag) { - return sycl::ext::oneapi::experimental::properties{sycl::ext::intel::experimental::streaming_interface<>, - sycl::ext::intel::experimental::pipelined<>}; + return sycl::ext::oneapi::experimental::properties{ +#ifdef AHLS + sycl::ext::altera::experimental::streaming_interface<>, sycl::ext::altera::experimental::pipelined<> +#else + sycl::ext::intel::experimental::streaming_interface<>, sycl::ext::intel::experimental::pipelined<> +#endif + }; } SYCL_EXTERNAL void operator()() const; diff --git a/hls4ml/templates/oneapi/firmware/nnet_utils/nnet_common.h b/hls4ml/templates/oneapi/firmware/nnet_utils/nnet_common.h index f37a61cb0c..7c6818e5dc 100644 --- a/hls4ml/templates/oneapi/firmware/nnet_utils/nnet_common.h +++ b/hls4ml/templates/oneapi/firmware/nnet_utils/nnet_common.h @@ -2,9 +2,16 @@ #define NNET_COMMON_H_ #include "nnet_helpers.h" + +#ifdef AHLS +#include +#include +#include +#else #include #include #include +#endif typedef ac_fixed<16, 6> table_default_t; diff --git a/hls4ml/templates/oneapi/myproject_bridge.cpp b/hls4ml/templates/oneapi/myproject_bridge.cpp index ddad1d054b..01468c64f5 100644 --- a/hls4ml/templates/oneapi/myproject_bridge.cpp +++ b/hls4ml/templates/oneapi/myproject_bridge.cpp @@ -56,7 +56,11 @@ void collect_trace_output(struct trace_data *c_trace_outputs) { void myproject_float( // hls-fpga-machine-learning insert header #float ) { +#ifdef AHLS + auto selector = sycl::ext::altera::fpga_emulator_selector_v; +#else auto selector = sycl::ext::intel::fpga_emulator_selector_v; +#endif static sycl::queue q(selector, fpga_tools::exception_handler, sycl::property::queue::enable_profiling{}); // hls-fpga-machine-learning insert wrapper #float @@ -67,7 +71,11 @@ void myproject_float( void myproject_double( // hls-fpga-machine-learning insert header #double ) { +#ifdef AHLS + auto selector = sycl::ext::altera::fpga_emulator_selector_v; +#else auto selector = sycl::ext::intel::fpga_emulator_selector_v; +#endif static sycl::queue q(selector, fpga_tools::exception_handler, sycl::property::queue::enable_profiling{}); // hls-fpga-machine-learning insert wrapper #double diff --git a/hls4ml/templates/oneapi/myproject_test.cpp b/hls4ml/templates/oneapi/myproject_test.cpp index 82fb60d2f8..a861e1fc69 100644 --- a/hls4ml/templates/oneapi/myproject_test.cpp +++ b/hls4ml/templates/oneapi/myproject_test.cpp @@ -9,9 +9,13 @@ #include "firmware/myproject.h" #include "firmware/parameters.h" +#ifdef AHLS +#include +#else #include +#endif -#if (__INTEL_CLANG_COMPILER < 20250000) +#if (__INTEL_CLANG_COMPILER < 20250000) && !defined(AHLS) #include #endif @@ -22,12 +26,22 @@ int main(int argc, char **argv) { +#ifdef AHLS +#if FPGA_SIMULATOR + auto selector = sycl::ext::altera::fpga_simulator_selector_v; +#elif FPGA_HARDWARE + auto selector = sycl::ext::altera::fpga_selector_v; +#else // #if FPGA_EMULATOR + auto selector = sycl::ext::altera::fpga_emulator_selector_v; +#endif +#else #if FPGA_SIMULATOR auto selector = sycl::ext::intel::fpga_simulator_selector_v; #elif FPGA_HARDWARE auto selector = sycl::ext::intel::fpga_selector_v; #else // #if FPGA_EMULATOR auto selector = sycl::ext::intel::fpga_emulator_selector_v; +#endif #endif sycl::queue q(selector, fpga_tools::exception_handler, sycl::property::queue::enable_profiling{});