Implement extension for dpnp.choose

[ndgrigorian]

This PR implements a new SYCL kernel for dpnp.choose in a new extension, _indexing_impl.

choose is an embarrassingly parallel copy operation, where the array to copy from is chosen from a list based on the index in the indices array.

• Have you provided a meaningful PR description?
• Have you added a test, reproducer or referred to issue with a reproducer?
• Have you tested your changes locally for CPU and GPU devices?
• Have you made sure that new changes do not introduce compiler warnings?
• Have you checked performance impact of proposed changes?
• If this PR is a work in progress, are you filing the PR as a draft?

[ndgrigorian]

Tests will need to be added in general, and especially to test_usm_type and test_sycl_queue, but suggestions are committed, as well as existing tests passing, so this PR is ready for first phase of review @antonwolfy

[coveralls]

coverage: 71.211% (+0.4%) from 70.856%
when pulling 925b7d8 on ndgrigorian:dedicated-choose-kernel
into 356184a on IntelPython:master.

[AlexanderKalistratov]

It doesn't feels right that keeping args alive depends on lifetime of temporaries.
I believe you are doing it in order to combine several independent events into one.

But this could potentially delay deletion of args resulting in increasing of memory consumption.

Though I don't know if it actually happens

[AlexanderKalistratov]

This function accepts too many parameters and it is too easy to get confused.

It also doing several things at a time going against single responsibility approach.

I'd refactor it into something like:

auto host_size_offset = make_size_offset();
auto host_chc_ptrs_shp = make_chc_ptrs();
auto host_shape_strides_shp = make_shape_strides();

std::vector<sycl::events> copy_evnts;
sycl::event host_event;
auto copy_evnts = batch_copy(exec_q, {device_chc_ptrs, host_chc_ptrs_shp}, {device_shape_strides, host_shape_strides_shp}, {device_chc_offsets, host_chc_offsets_shp});
auto host_evnt = async_free(exec_q, copy_evnts, host_size_offset, host_chc_ptrs_shp, host_shape_strides_shp)

Where batch_copy is some hypothetical variadic function which enques copy.