[PATCH] D60455: [SYCL] Add support for SYCL device attributes

[Anastasia Stulova via Phabricator via cfe-commits]

Anastasia added a comment.



> Our current approach is to add an attribute, which SYCL runtime will use to mark code passed to `cl::sycl::handler::parallel_for` as "kernel functions". Obviously runtime library can't mark `foo` as "device" code - this is a compiler job: to traverse all symbols accessible from kernel functions and add them to the "device part" of the code.
> 
> Here is a link to the code in the SYCL runtime using `sycl_kernel` attribute: https://github.com/intel/llvm/blob/sycl/sycl/include/CL/sycl/handler.hpp#L267
> 
> I'm quite sure something similar should happen for other "single source" programming models like OpenMP/CUDA, except these attributes are exposed to the user and there is a specific requirement on attributes/pragma/keyword names.

Just to understand how this will work. I would imagine you can have a device function definition preceding its use. When the function is being parsed it's not known yet whether it will be called from the device or not, so it won't be possible to set the language mode correctly and hence provide the right diagnostics. So is the plan to launch a separate parsing phase then just to extract the call graph and annotate the device functions?

> NOTE2: @Anastasia, https://reviews.llvm.org/D60454 makes impossible to replace `sycl_kernel` attribute with `__kernel` attribute. I mean we still can enable it for SYCL extension, but we will need SYCL specific customization in this case as we apply `kernel` attribute to a template function.

Ok, my question is whether you are planning to duplicate the same logic as for OpenCL kernel which doesn't really seem like an ideal design choice. Is this the only difference then we can simply add an extra check for SYCL compilation mode in this template handling case. The overall interaction between OpenCL and SYCL implementation is still a very big unknown to me so it's not very easy to judge about the implementations details...


Repository:
  rG LLVM Github Monorepo

CHANGES SINCE LAST ACTION
  https://reviews.llvm.org/D60455/new/

https://reviews.llvm.org/D60455