Fall BackΒΆ

Typically, a command group is submitted and executed on the designated command queue; however, there may be cases where the command queue is unable to execute the group. In these cases, it is possible to specify a fall back command queue for the command group to be executed upon. This capability is handled by the runtime. This fallback mechanism is detailed in the SYCL Specification.

The following code fails due to the size of the workgroup when executed on Intel Processor Graphics, such as Intel HD Graphics 530. The SYCL specification allows specifying a secondary queue as a parameter to the submit function and this secondary queue is used if the device kernel runs into issues with submission to the first device.

 1#include<CL/sycl.hpp>
 2#include<iostream>
 3
 4const int N = 1024;
 5const int M = 32;
 6
 7using namespace sycl;
 8
 9int main(){
10  cpu_selector cpuSelector;
11  queue cpuQueue(cpuSelector);
12  queue defaultqueue;
13  buffer<int,2> buf(range<2>(N,N));
14
15  defaultqueue.submit([&](handler &h){
16      auto bufacc = buf.get_access<access::mode::read_write>(h);
17      h.parallel_for(nd_range<2>(range<2>(N,N), range<2>(M,M)),
18		     [=](nd_item<2> i){
19		       id<2> ind = i.get_global_id();
20		       bufacc[ind[0]][ind[1]] = ind[0]+ind[1];
21		     });
22    }, cpuQueue);
23  auto bufacc1 = buf.get_access<access::mode::read>();
24  for(int i = 0; i < N; i++){
25    for(int j = 0; j < N; j++){
26      if(bufacc1[i][j] != i+j){
27	std::cout<<"Wrong result\n";
28	return 1;
29      }
30    }
31  }
32  std::cout<<"Correct results\n";
33  return 0;
34}

See also