FPGA: Improve the fpga_compile tutorial

DirectProgramming/C++SYCL_FPGA/Tutorials/GettingStarted/fpga_compile/CMakeLists.txt → DirectProgramming/C++SYCL_FPGA/Tutorials/GettingStarted/fpga_compile/part1-C++/CMakeLists.txt

@@ -9,12 +9,12 @@ else() # Windows
     include (Platform/Windows-Clang)
 endif()
 
-cmake_minimum_required (VERSION 3.4)
+cmake_minimum_required (VERSION 3.7.2)
 
 project(FPGACompile CXX)
 
 set(CMAKE_ARCHIVE_OUTPUT_DIRECTORY ${CMAKE_BINARY_DIR})
 set(CMAKE_LIBRARY_OUTPUT_DIRECTORY ${CMAKE_BINARY_DIR})
 set(CMAKE_RUNTIME_OUTPUT_DIRECTORY ${CMAKE_BINARY_DIR})
 
-add_subdirectory (src)
+add_subdirectory (src)

DirectProgramming/C++SYCL_FPGA/Tutorials/GettingStarted/fpga_compile/part1-C++/src/CMakeLists.txt

@@ -0,0 +1,25 @@
+
+set(SOURCE_FILE vector_add.cpp)
+set(TARGET_NAME vector_add)
+
+# FPGA device selection
+if(DEFINED FPGA_DEVICE)
+    message(STATUS "Ignoring FPGA_DEVICE: ${FPGA_DEVICE}, not applicable")
+endif()
+
+set(EMULATOR_TARGET ${TARGET_NAME}.fpga_emu)
+
+add_executable(${EMULATOR_TARGET} ${SOURCE_FILE})
+add_custom_target(fpga_emu DEPENDS ${EMULATOR_TARGET})
+
+# This code sample do not support simulator and fpga. 
+# following targets are added to be compatiable with reg-tests
+set(SIMULATOR_TARGET ${TARGET_NAME}.fpga_sim)
+
+add_executable(${SIMULATOR_TARGET} ${SOURCE_FILE})
+add_custom_target(fpga_sim DEPENDS ${SIMULATOR_TARGET})
+
+set(FPGA_TARGET ${TARGET_NAME}.fpga)
+
+add_executable(${FPGA_TARGET} ${SOURCE_FILE})
+add_custom_target(fpga DEPENDS ${FPGA_TARGET})

DirectProgramming/C++SYCL_FPGA/Tutorials/GettingStarted/fpga_compile/part1-C++/src/vector_add.cpp

@@ -0,0 +1,46 @@
+#include <iostream>
+
+void VectorAdd(const int *a_in, const int *b_in, int *c_out, int len) {
+    for (int idx = 0; idx < len; idx++) {
+        int a_val  = a_in[idx];
+        int b_val  = b_in[idx];
+        int sum    = a_val + b_val;
+        c_out[idx] = sum;
+    }
+}
+
+constexpr int kVectSize = 256;
+
+int main() {
+
+    // declare arrays and fill them
+    int *vec_a = new int[kVectSize];
+    int *vec_b = new int[kVectSize];
+    int *vec_c = new int[kVectSize];
+    for (int i = 0; i < kVectSize; i++) {
+        vec_a[i] = i;
+        vec_b[i] = (kVectSize - i);
+    }
+
+    std::cout << "add two vectors of size " << kVectSize << std::endl;
+
+    VectorAdd(vec_a, vec_b, vec_c, kVectSize);
+
+    // verify that vector C is correct
+    bool passed = true;
+    for (int i = 0; i < kVectSize; i++) {
+        int expected = vec_a[i] + vec_b[i];
+        if (vec_c[i] != expected) {
+            std::cout << "idx=" << i << ": result " << vec_c[i] << ", expected (" << expected << ") A=" << vec_a[i] << " + B=" << vec_b[i] << std::endl;
+            passed = false;
+        }
+    }
+
+    std::cout << (passed ? "PASSED" : "FAILED") << std::endl;
+
+    delete[] vec_a;
+    delete[] vec_b;
+    delete[] vec_c;
+
+    return passed ? EXIT_SUCCESS : EXIT_FAILURE;
+}

DirectProgramming/C++SYCL_FPGA/Tutorials/GettingStarted/fpga_compile/part2-dpcpp_functor_usm/CMakeLists.txt

@@ -0,0 +1,20 @@
+if(UNIX)
+    # Direct CMake to use dpcpp rather than the default C++ compiler/linker
+    set(CMAKE_CXX_COMPILER icpx)
+else() # Windows
+    # Force CMake to use dpcpp rather than the default C++ compiler/linker 
+    # (needed on Windows only)
+    include (CMakeForceCompiler)
+    CMAKE_FORCE_CXX_COMPILER (icx-cl IntelDPCPP)
+    include (Platform/Windows-Clang)
+endif()
+
+cmake_minimum_required (VERSION 3.7.2)
+
+project(FPGACompile CXX)
+
+set(CMAKE_ARCHIVE_OUTPUT_DIRECTORY ${CMAKE_BINARY_DIR})
+set(CMAKE_LIBRARY_OUTPUT_DIRECTORY ${CMAKE_BINARY_DIR})
+set(CMAKE_RUNTIME_OUTPUT_DIRECTORY ${CMAKE_BINARY_DIR})
+
+add_subdirectory (src)

DirectProgramming/C++SYCL_FPGA/Tutorials/GettingStarted/fpga_compile/src/CMakeLists.txt → DirectProgramming/C++SYCL_FPGA/Tutorials/GettingStarted/fpga_compile/part2-dpcpp_functor_usm/src/CMakeLists.txt

@@ -1,5 +1,5 @@
-set(SOURCE_FILE fpga_compile.cpp)
-set(TARGET_NAME fpga_compile)
+set(SOURCE_FILE vector_add.cpp)
+set(TARGET_NAME vector_add)
 set(EMULATOR_TARGET ${TARGET_NAME}.fpga_emu)
 set(SIMULATOR_TARGET ${TARGET_NAME}.fpga_sim)
 set(FPGA_TARGET ${TARGET_NAME}.fpga)
@@ -14,77 +14,86 @@ else()
     message(STATUS "Configuring the design with the following target: ${FPGA_DEVICE}")
 endif()
 
-# This is a Windows-specific flag that enables exception handling in host code
+# These are Windows-specific flags:
+# 1. /EHsc This is a Windows-specific flag that enables exception handling in host code
+# 2. /Qactypes Include ac_types headers and link against ac_types emulation libraries
 if(WIN32)
     set(WIN_FLAG "/EHsc")
+    set(AC_TYPES_FLAG "/Qactypes")
+else()
+    set(AC_TYPES_FLAG "-qactypes")
 endif()
 
 # A SYCL ahead-of-time (AoT) compile processes the device code in two stages.
 # 1. The "compile" stage compiles the device code to an intermediate representation (SPIR-V).
 # 2. The "link" stage invokes the compiler's FPGA backend before linking.
 #    For this reason, FPGA backend flags must be passed as link flags in CMake.
-set(EMULATOR_COMPILE_FLAGS "-fsycl -fintelfpga -Wall ${WIN_FLAG} -DFPGA_EMULATOR")
-set(EMULATOR_LINK_FLAGS "-fsycl -fintelfpga")
-set(SIMULATOR_COMPILE_FLAGS "-fsycl -fintelfpga -Wall ${WIN_FLAG} -Xssimulation -DFPGA_SIMULATOR")
-set(SIMULATOR_LINK_FLAGS "-fsycl -fintelfpga -Xssimulation -Xsghdl -Xstarget=${FPGA_DEVICE} ${USER_HARDWARE_FLAGS}")
-set(HARDWARE_COMPILE_FLAGS "-fsycl -fintelfpga -Wall ${WIN_FLAG} -DFPGA_HARDWARE")
-set(HARDWARE_LINK_FLAGS "-fsycl -fintelfpga -Xshardware -Xstarget=${FPGA_DEVICE} ${USER_HARDWARE_FLAGS}")
+set(EMULATOR_COMPILE_FLAGS "-fsycl -fintelfpga ${AC_TYPES_FLAG} -DFPGA_EMULATOR -Wall ${WIN_FLAG}")
+set(EMULATOR_LINK_FLAGS "-fsycl -fintelfpga ${AC_TYPES_FLAG}")
+set(SIMULATOR_COMPILE_FLAGS "-fsycl -fintelfpga ${AC_TYPES_FLAG} -DFPGA_SIMULATOR -Wall ${WIN_FLAG}")
+set(SIMULATOR_LINK_FLAGS "-fsycl -fintelfpga ${AC_TYPES_FLAG} -Xssimulation -Xsghdl -Xstarget=${FPGA_DEVICE} ${USER_HARDWARE_FLAGS}")
+set(REPORT_COMPILE_FLAGS "-fsycl -fintelfpga ${AC_TYPES_FLAG} -Wall ${WIN_FLAG} -DFPGA_REPORT")
+set(REPORT_LINK_FLAGS "-fsycl -fintelfpga -Xshardware -Xstarget=${FPGA_DEVICE} ${USER_HARDWARE_FLAGS}")
+set(HARDWARE_COMPILE_FLAGS "-fsycl -fintelfpga ${AC_TYPES_FLAG} -Wall ${WIN_FLAG} -DFPGA_HARDWARE")
+set(HARDWARE_LINK_FLAGS "-fsycl -fintelfpga ${AC_TYPES_FLAG} -Xshardware -Xstarget=${FPGA_DEVICE} ${USER_HARDWARE_FLAGS}")
 # use cmake -D USER_HARDWARE_FLAGS=<flags> to set extra flags for FPGA backend compilation
 
 ###############################################################################
 ### FPGA Emulator
 ###############################################################################
 # To compile in a single command:
-#    icpx -fsycl -fintelfpga -DFPGA_EMULATOR fpga_compile.cpp -o fpga_compile.fpga_emu
+#    icpx -fsycl -fintelfpga ${AC_TYPES_FLAG} -DFPGA_EMULATOR fpga_compile.cpp -o fpga_compile.fpga_emu
 # CMake executes:
-#    [compile] icpx -fsycl -fintelfpga -DFPGA_EMULATOR -o fpga_compile.cpp.o -c fpga_compile.cpp
-#    [link]    icpx -fsycl -fintelfpga fpga_compile.cpp.o -o fpga_compile.fpga_emu
+#    [compile] icpx -fsycl -fintelfpga ${AC_TYPES_FLAG} -DFPGA_EMULATOR -o fpga_compile.cpp.o -c fpga_compile.cpp
+#    [link]    icpx -fsycl -fintelfpga ${AC_TYPES_FLAG} fpga_compile.cpp.o -o fpga_compile.fpga_emu
 add_executable(${EMULATOR_TARGET} ${SOURCE_FILE})
 target_include_directories(${EMULATOR_TARGET} PRIVATE ../../../../include)
 set_target_properties(${EMULATOR_TARGET} PROPERTIES COMPILE_FLAGS "${EMULATOR_COMPILE_FLAGS}")
 set_target_properties(${EMULATOR_TARGET} PROPERTIES LINK_FLAGS "${EMULATOR_LINK_FLAGS}")
 add_custom_target(fpga_emu DEPENDS ${EMULATOR_TARGET})
 
-###############################################################################
-### FPGA Simulator
-###############################################################################
-# To compile in a single command:
-#    icpx -fsycl -fintelfpga -Xssimulation -Xsghdl -Xstarget=<FPGA_DEVICE> -DFPGA_SIMULATOR <file>.cpp -o <file>.fpga_sim
-# CMake executes:
-#    [compile] icpx -fsycl -fintelfpga -Xssimulation -DFPGA_SIMULATOR -o <file>.cpp.o -c <file>.cpp
-#    [link]    icpx -fsycl -fintelfpga -Xssimulation -Xsghdl -Xstarget=<FPGA_DEVICE> <file>.cpp.o -o <file>.fpga_sim
-add_executable(${SIMULATOR_TARGET} ${SOURCE_FILE})
-target_include_directories(${SIMULATOR_TARGET} PRIVATE ../../../../include)
-set_target_properties(${SIMULATOR_TARGET} PROPERTIES COMPILE_FLAGS "${SIMULATOR_COMPILE_FLAGS}")
-set_target_properties(${SIMULATOR_TARGET} PROPERTIES LINK_FLAGS "${SIMULATOR_LINK_FLAGS}")
-add_custom_target(fpga_sim DEPENDS ${SIMULATOR_TARGET})
-
 ###############################################################################
 ### Generate Report
 ###############################################################################
 # To compile manually:
-#   icpx -fsycl -fintelfpga -Xshardware -Xstarget=<FPGA_DEVICE> -fsycl-link=early fpga_compile.cpp -o fpga_compile_report.a
+#   icpx -fsycl -fintelfpga ${AC_TYPES_FLAG} -Xshardware -Xstarget=<FPGA_DEVICE> -fsycl-link=early ac_fixed.cpp -o ac_fixed_report.a
 set(FPGA_EARLY_IMAGE ${TARGET_NAME}_report.a)
 # The compile output is not an executable, but an intermediate compilation result unique to SYCL.
 add_executable(${FPGA_EARLY_IMAGE} ${SOURCE_FILE})
 target_include_directories(${FPGA_EARLY_IMAGE} PRIVATE ../../../../include)
 add_custom_target(report DEPENDS ${FPGA_EARLY_IMAGE})
-set_target_properties(${FPGA_EARLY_IMAGE} PROPERTIES COMPILE_FLAGS "${HARDWARE_COMPILE_FLAGS}")
-set_target_properties(${FPGA_EARLY_IMAGE} PROPERTIES LINK_FLAGS "${HARDWARE_LINK_FLAGS} -fsycl-link=early")
+set_target_properties(${FPGA_EARLY_IMAGE} PROPERTIES COMPILE_FLAGS "${REPORT_COMPILE_FLAGS}")
+set_target_properties(${FPGA_EARLY_IMAGE} PROPERTIES LINK_FLAGS "${REPORT_LINK_FLAGS} -fsycl-link=early")
 # fsycl-link=early stops the compiler after RTL generation, before invoking Quartus®
 
+###############################################################################
+### FPGA Simulator
+###############################################################################
+# To compile in a single command:
+#   icpx -fsycl -fintelfpga -DFPGA_SIMULATOR ${AC_TYPES_FLAG} -Xssimulation -Xsghdl -Xstarget=<FPGA_DEVICE> ac_fixed.cpp -o ac_fixed.fpga
+# CMake executes:
+#   [compile] icpx -fsycl -fintelfpga -DFPGA_SIMULATOR ${AC_TYPES_FLAG} -o ac_fixed.cpp.o -c ac_fixed.cpp
+#   [link]    icpx -fsycl -fintelfpga ${AC_TYPES_FLAG} -Xssimulation -Xsghdl -Xstarget=<FPGA_DEVICE> ac_fixed.cpp.o -o ac_fixed.fpga
+add_executable(${SIMULATOR_TARGET} EXCLUDE_FROM_ALL ${SOURCE_FILE})
+target_include_directories(${SIMULATOR_TARGET} PRIVATE ../../../../include)
+add_custom_target(fpga_sim DEPENDS ${SIMULATOR_TARGET})
+set_target_properties(${SIMULATOR_TARGET} PROPERTIES COMPILE_FLAGS "${SIMULATOR_COMPILE_FLAGS}")
+set_target_properties(${SIMULATOR_TARGET} PROPERTIES LINK_FLAGS "${SIMULATOR_LINK_FLAGS} -reuse-exe=${CMAKE_BINARY_DIR}/${SIMULATOR_TARGET}")
+# The -reuse-exe flag enables rapid recompilation of host-only code changes.
+# See C++SYCL_FPGA/GettingStarted/fast_recompile for details.
+
 ###############################################################################
 ### FPGA Hardware
 ###############################################################################
 # To compile in a single command:
-#   icpx -fsycl -fintelfpga -Xshardware -Xstarget=<FPGA_DEVICE> fpga_compile.cpp -o fpga_compile.fpga
+#   icpx -fsycl -fintelfpga ${AC_TYPES_FLAG} -Xshardware -Xstarget=<FPGA_DEVICE> ac_fixed.cpp -o ac_fixed.fpga
 # CMake executes:
-#   [compile] icpx -fsycl -fintelfpga -o fpga_compile.cpp.o -c fpga_compile.cpp
-#   [link]    icpx -fsycl -fintelfpga -Xshardware -Xstarget=<FPGA_DEVICE> fpga_compile.cpp.o -o fpga_compile.fpga
+#   [compile] icpx -fsycl -fintelfpga ${AC_TYPES_FLAG} -o ac_fixed.cpp.o -c ac_fixed.cpp
+#   [link]    icpx -fsycl -fintelfpga ${AC_TYPES_FLAG} -Xshardware -Xstarget=<FPGA_DEVICE> ac_fixed.cpp.o -o ac_fixed.fpga
 add_executable(${FPGA_TARGET} EXCLUDE_FROM_ALL ${SOURCE_FILE})
 target_include_directories(${FPGA_TARGET} PRIVATE ../../../../include)
 add_custom_target(fpga DEPENDS ${FPGA_TARGET})
 set_target_properties(${FPGA_TARGET} PROPERTIES COMPILE_FLAGS "${HARDWARE_COMPILE_FLAGS}")
-set_target_properties(${FPGA_TARGET} PROPERTIES LINK_FLAGS "${HARDWARE_LINK_FLAGS}")
-
-
+set_target_properties(${FPGA_TARGET} PROPERTIES LINK_FLAGS "${HARDWARE_LINK_FLAGS} -reuse-exe=${CMAKE_BINARY_DIR}/${FPGA_TARGET}")
+# The -reuse-exe flag enables rapid recompilation of host-only code changes.
+# See C++SYCL_FPGA/GettingStarted/fast_recompile for details.

DirectProgramming/C++SYCL_FPGA/Tutorials/GettingStarted/fpga_compile/part2-dpcpp_functor_usm/src/vector_add.cpp

@@ -0,0 +1,97 @@
+#include <iostream>
+
+// oneAPI headers
+#include <sycl/sycl.hpp>
+#include <sycl/ext/intel/fpga_extensions.hpp>
+
+// Forward declare the kernel name in the global scope. This is an FPGA best
+// practice that reduces name mangling in the optimization reports.
+class VectorAddID;
+
+struct VectorAdd {
+  public:
+    int *const vec_a_in;
+    int *const vec_b_in;
+    int *const vec_c_out;
+    int len;
+
+    void operator()() const {
+        for (int idx = 0; idx < len; idx++) {
+            int a_val  = vec_a_in[idx];
+            int b_val  = vec_b_in[idx];
+            int sum    = a_val + b_val;
+            vec_c_out[idx] = sum;
+        }
+    }
+};
+
+constexpr int kVectSize = 256;
+
+int main() {
+    bool passed = true;
+    try {
+        // Use compile-time macros to select either:
+        //  - the FPGA emulator device (CPU emulation of the FPGA)
+        //  - the FPGA device (a real FPGA)
+        //  - the simulator device 
+#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
+
+        // create the device queue
+        sycl::queue q(selector);
+
+        auto device = q.get_device();
+
+        std::cout << "Running on device: "
+                << device.get_info<sycl::info::device::name>().c_str()
+                << std::endl;
+
+        // declare arrays and fill them
+        // allocate in shared memory so the kernel can see them
+        int *vec_a = sycl::malloc_shared<int>(kVectSize, q);
+        int *vec_b = sycl::malloc_shared<int>(kVectSize, q);
+        int *vec_c = sycl::malloc_shared<int>(kVectSize, q);
+        for (int i = 0; i < kVectSize; i++) {
+            vec_a[i] = i;
+            vec_b[i] = (kVectSize - i);
+        }
+
+        std::cout << "add two vectors of size " << kVectSize << std::endl;
+
+        q.single_task<VectorAddID>(VectorAdd{vec_a, vec_b, vec_c, kVectSize}).wait();
+
+        // verify that vec_c is correct
+        for (int i = 0; i < kVectSize; i++) {
+            int expected = vec_a[i] + vec_b[i];
+            if (vec_c[i] != expected) {
+                std::cout << "idx=" << i << ": result " << vec_c[i] << ", expected (" << expected << ") A=" << vec_a[i] << " + B=" << vec_b[i] << std::endl;
+                passed = false;
+            }
+        }
+
+        std::cout << (passed ? "PASSED" : "FAILED") << std::endl;
+
+        sycl::free(vec_a, q);
+        sycl::free(vec_b, q);
+        sycl::free(vec_c, q);
+    } catch (sycl::exception const &e) {
+        // Catches exceptions in the host code.
+        std::cerr << "Caught a SYCL host exception:\n" << e.what() << "\n";
+
+        // Most likely the runtime couldn't find FPGA hardware!
+        if (e.code().value() == CL_DEVICE_NOT_FOUND) {
+        std::cerr << "If you are targeting an FPGA, please ensure that your "
+                    "system has a correctly configured FPGA board.\n";
+        std::cerr << "Run sys_check in the oneAPI root directory to verify.\n";
+        std::cerr << "If you are targeting the FPGA emulator, compile with "
+                    "-DFPGA_EMULATOR.\n";
+        }
+        std::terminate();
+    }
+    return passed ? EXIT_SUCCESS : EXIT_FAILURE;
+}

DirectProgramming/C++SYCL_FPGA/Tutorials/GettingStarted/fpga_compile/part3-dpcpp_lambda_usm/CMakeLists.txt

@@ -0,0 +1,20 @@
+if(UNIX)
+    # Direct CMake to use dpcpp rather than the default C++ compiler/linker
+    set(CMAKE_CXX_COMPILER icpx)
+else() # Windows
+    # Force CMake to use dpcpp rather than the default C++ compiler/linker 
+    # (needed on Windows only)
+    include (CMakeForceCompiler)
+    CMAKE_FORCE_CXX_COMPILER (icx-cl IntelDPCPP)
+    include (Platform/Windows-Clang)
+endif()
+
+cmake_minimum_required (VERSION 3.7.2)
+
+project(FPGACompile CXX)
+
+set(CMAKE_ARCHIVE_OUTPUT_DIRECTORY ${CMAKE_BINARY_DIR})
+set(CMAKE_LIBRARY_OUTPUT_DIRECTORY ${CMAKE_BINARY_DIR})
+set(CMAKE_RUNTIME_OUTPUT_DIRECTORY ${CMAKE_BINARY_DIR})
+
+add_subdirectory (src)