Suggested Kernel Coding Styles

Developer Guide

Intel® oneAPI DPC++/C++ Compiler Handbook for FPGAs

Download PDF

ID 785441

Date 10/24/2024

Version

Public

Visible to Intel only — GUID: GUID-8EAA57D0-08FB-4A2B-B2AD-5737A6B9F117

View Details

Suggested Kernel Coding Styles

For creating your kernel, use one of the following recommended general coding styles:

Functor Coding Style Example: With the functor coding style, you can write your kernel code out-of-line from the host code. The style is typically preferred for the SYCL* HLS flow. In the SYCL* HLS flow, your host code becomes the testbench for your IP core.
The functor coding style clearly communicates what the inputs to your kernel are and you can easily annotate kernel arguments.
Lambda Coding Style Example: The lambda coding style is typically used for kernels in the FPGA acceleration flow.

In the following examples, the kernel code and the code for its enqueueing is highlighted.

Functor Coding Style Example

#include <iostream>

// oneAPI headers
#include <sycl/ext/intel/fpga_extensions.hpp>
#include <sycl/sycl.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 {
  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);

   // make sure the device supports USM host allocations
    auto device = q.get_device();

    std::cout << "Running on device: "
              << device.get_info<sycl::info::device::name>().c_str()
              << std::endl;

    if (!device.has(sycl::aspect::usm_host_allocations)) {
      std::terminate();
    }

    // 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);
    assert(vec_a);
    assert(vec_b);
    assert(vec_c);
    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;
}

Lambda Coding Style Example

 #include <iostream>
 #include <vector>

 // oneAPI headers
 #include <sycl/sycl.hpp>
 #include <sycl/ext/intel/fpga_extensions.hpp>

 using namespace sycl;

 // 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;

 void VectorAdd(const int *vec_a_in, const int *vec_b_in, int *vec_c_out,
                int len) {
   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);

     // make sure the device supports USM host allocations
     auto device = q.get_device();

     std::cout << "Running on device: "
               << device.get_info<sycl::info::device::name>().c_str()
               << std::endl;

     if (!device.has(sycl::aspect::usm_host_allocations)) {
        std::terminate();
     }

     // declare arrays and fill them
     // allocate in shared memory so the kernel can see them
     int *vec_a = malloc_shared<int>(kVectSize, q);
     int *vec_b = malloc_shared<int>(kVectSize, q);
     int *vec_c = malloc_shared<int>(kVectSize, q);
     assert(vec_a);
     assert(vec_b);
     assert(vec_c);
     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;

     free(vec_a, q);
     free(vec_b, q);
     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;
 }

Select Your Language

Using Intel.com Search

Quick Links

Recent Searches

Advanced Search

Only search in

Intel® oneAPI DPC++/C++ Compiler Handbook for FPGAs

Suggested Kernel Coding Styles

Functor Coding Style Example

Lambda Coding Style Example