topic Re:Request for Intel OpenCL Offline Compiler (OCLOC) in Intel® oneAPI DPC++/C++ Compiler

Request for Intel OpenCL Offline Compiler (OCLOC)

Viet-Duc — Fri, 28 May 2021 04:51:46 GMT

Hi,

I am trying to perform ahead of time compilation following the instruction here:

https://software.intel.com/content/www/us/en/develop/documentation/oneapi-dpcpp-cpp-compiler-dev-guide-and-reference/top/compilation/ahead-of-time-compilation.html

https://software.intel.com/content/www/us/en/develop/documentation/installation-guide-for-intel-oneapi-toolkits-linux/top/install-opencl-offline-compiler-ocloc.html

Using the following compiler options:

dpcpp -fsycl-targets=spir64_gen-unknown-unknown-sycldevice -Xs "-device Gen9,dg1,Gen12HP" vector-add.cpp.

The error message is:

dpcpp: error: unable to execute command: Executable "ocloc" doesn't exist!

As I understand, the NDA queue is not configured to accept interactive job (qsub -I).

If I compile on login node and run on Xe_HP, the performance is not as good as expected. For instance, the result of memory bandwidth measurement is aproximately 40% of theoretical maximum of 800GB/s.

Would you make ocloc available system-wide ? If not, is there a way to install it locally ?

Regards.

Re:Request for Intel OpenCL Offline Compiler (OCLOC)

Gopika_Intel — Fri, 28 May 2021 11:53:14 GMT

Hi,

Thank you for reaching out. We are checking this internally. We will get back to you as soon as we get an update.

Regards

Gopika

Re: Request for Intel OpenCL Offline Compiler (OCLOC)

Viet-Duc — Sat, 29 May 2021 06:49:17 GMT

Thanks for moving the post to the suitable forum.

I am not sure if I can show performance of NDA hardware here. Nevertheless, I need to make my case.

The measurement is done using BabelStream, which is a extension of Stream benchmark for heterogenous platforms.

The triad kernel implemented in SYCL is as follow:

template <class T> void SYCLStream<T>::triad() { const T scalar = startScalar; queue->submit([&](handler &cgh) { auto ka = d_a->template get_access<access::mode::write>(cgh); auto kb = d_b->template get_access<access::mode::read>(cgh); auto kc = d_c->template get_access<access::mode::read>(cgh); cgh.parallel_for<triad_kernel>(range<1>{array_size}, [=](id<1> idx) { ka[idx] = kb[idx] + scalar * kc[idx]; }); }); queue->wait(); }

This implementation does not use USM and work group size is determined by OpenCL runtime, which may affect performance.

Result for DG1 is below:

Array size: 268.4 MB (=0.3 GB) Total size: 805.3 MB (=0.8 GB) Using SYCL device Intel(R) Iris(R) Xe MAX Graphics [0x4905] Driver: 21.11.19310 Function MBytes/sec Min (sec) Max Average Triad 59800.403 0.01347 0.01361 0.01354

This translates to 86% of theoretical bandwidth.

Result for Xe_HP is below:

... Using SYCL device Intel(R) Graphics [0x0205] Driver: 21.12.019357+embargo458 Function MBytes/sec Min (sec) Max Averag Triad 282888.513 0.00285 0.00368 0.00345

This translates to only 36% of the theoretical bandwidth. Tuning workgroup size may give 10% margin gain.

Thus, I am trying to use AOT compilation since the NDA queue does not accept iteractive jobs.

Your insights on this issue is much appreciated.

Thanks.

Re:Request for Intel OpenCL Offline Compiler (OCLOC)

Jie_L_Intel — Wed, 02 Jun 2021 11:08:19 GMT

The devcloud does not install the ocloc right now. You could try it following the installation guide if you have your own develop machine that could install oneAPI.

https://software.intel.com/content/www/us/en/develop/documentation/installation-guide-for-intel-oneapi-toolkits-linux/top/installation/install-opencl-offline-compiler-ocloc.html