[SYCL] Hide SYCL service kernels

[steffenlarsen]

The SYCL runtime may in places benefit from being able to define "service kernels" to use behind-the-scenes work on devices. These kernels may use the same interface as other SYCL kernels, but as such will be handled no different than user-defined kernels.

These changes makes a distinction between service kernels and user-defined kernels by placing service kernels in a common namespace (cl::sycl::detail::__sycl_service_kernel__). The program manager will not grant unique kernel IDs to service kernels. This prevents service kernels from being visible to the user through kernel_bundle related interfaces.

Consequently the runtime may create device images that do not contain any kernel IDs. As such, sycl::has_kernel_bundle will now discount any device images that have no kernel IDs.

[s-kanaev]

This is quite straightforward approach and will work.
Though, I'm considered on its performance.
Is possible to get notion of mangling and check just the beginning of the string for Linux and ending for Windows?

[steffenlarsen]

Checking the start of the string was the original approach, but it did not consider Windows mangling. I am not very familiar with the Windows mangling scheme, but I worry that even if we could determine which mangling scheme is used, having templated classes would break the assumption about the end of the mangled name.

[romanovvlad]

I thought we have the same(Itanium) mangling for kernels on both windows and linux.
In any case, if my another suggest is applied, the only place where we check if a kernel is a service one would be addImages function which should be called once per process for a given image. Having said that I think this is not a performance critical peace of code.

[s-kanaev]

It seems, like my yesterday's comment didn't get to post here:
Is it possible to cache this flag?
I believe, using unordered_map will allow for quick access. Also, we can eliminate use of mutexes as the initialization is only needed at boot-strap and there are only read requests during run-time.

[steffenlarsen]

I believe, using unordered_map will allow for quick access.

You're thinking adding an unordered_set (I think set would do the trick) where AddImages caches all found service kernel names. Then ProgramManager::getSYCLDeviceImagesWithCompatibleState would check against that rather than calling isServiceKernel?

Also, we can eliminate use of mutexes as the initialization is only needed at boot-strap and there are only read requests during run-time.

Are you referring to m_KernelIDsMutex?

[s-kanaev]

You're thinking adding an unordered_set (I think set would do the trick) where AddImages caches all found service kernel names. Then ProgramManager::getSYCLDeviceImagesWithCompatibleState would check against that rather than calling isServiceKernel?

Right. This will do the thing.

Are you referring to m_KernelIDsMutex?

Never, never, never. I meant any mutex required for lazy initialization of cache. With the cache initialized at boot-strap there is no place for lazy initialization and, thus, no need for thread-safety mechanism except for ensuring that only read access takes place during run-time.

[steffenlarsen]

Right. This will do the thing.

I have implemented caching of the service kernels found in ProgramManager::AddImages. This allows for faster lookup later. However, one concern I have is that the only time later that service kernels are looked up is in an assert, so in release builds we will never look up anything in the new cache.

Never, never, never. I meant any mutex required for lazy initialization of cache. With the cache initialized at boot-strap there is no place for lazy initialization and, thus, no need for thread-safety mechanism except for ensuring that only read access takes place during run-time.

I think this may apply to m_KernelIDsMutex as well. That is outside the scope of this PR, so I'll make another PR in which we can discuss further.

[romanovvlad]

I thought we have the same(Itanium) mangling for kernels on both windows and linux.
In any case, if my another suggest is applied, the only place where we check if a kernel is a service one would be addImages function which should be called once per process for a given image. Having said that I think this is not a performance critical peace of code.