DPCT1026#

Message#

The call to <API name> was removed because <reason>.

Detailed Help#

API calls from the original application, which do not have functionally compatible SYCL* API calls are removed if SYCLomatic determines that it should not affect the program logic.

Possible reasons for removal:

  • SYCL currently does not support setting resource limits on devices.

  • SYCL currently does not support associating USM with a specific queue.

  • SYCL currently does not support query operations on queues.

  • SYCL currently does not support capture operations on queues.

  • SYCL currently does not support configuring shared memory on devices.

  • SYCL currently does not support setting cache config on devices.

  • SYCL currently does not support registering of existing host memory for use by device. Use USM to allocate memory for use by host and device.

  • SYCL currently does not support setting flags for devices.

  • SYCL currently does not support memory access across peer devices.

  • There is no corresponding API in SYCL.

  • The call is redundant in SYCL.

Suggestions to Fix#

Verify the code correctness.

For example, this original CUDA* code:

 1__device__ void bar(int *a) {
 2  __ldg(a);
 3}
 4
 5__global__ void kernel() {
 6  return;
 7}
 8
 9void foo() {
10  cudaLimit limit;
11  cudaStream_t stream;
12  unsigned int flags;
13  cudaSharedMemConfig config;
14
15  cudaDeviceSetLimit(limit, 0);
16  cudaStreamAttachMemAsync(stream, nullptr);
17  cudaStreamQuery(stream);
18  cudaDeviceSetSharedMemConfig(config);
19  cudaDeviceSetCacheConfig(cudaFuncCachePreferNone);
20  cudaSetDeviceFlags(flags);
21  cuInit(0);
22  float *aptr;
23  cudaStreamAttrValue stream_attribute = {};
24  stream_attribute.accessPolicyWindow.base_ptr = aptr;
25  stream_attribute.accessPolicyWindow.num_bytes = 8 * 8 * 8;
26  stream_attribute.accessPolicyWindow.hitRatio = 1.0f;
27  stream_attribute.accessPolicyWindow.hitProp = cudaAccessPropertyPersisting;
28  stream_attribute.accessPolicyWindow.missProp = cudaAccessPropertyStreaming;
29  cudaStreamSetAttribute(stream, cudaStreamAttributeAccessPolicyWindow, &stream_attribute);
30  cudaStreamGetAttribute(stream, cudaStreamAttributeAccessPolicyWindow, &stream_attribute);
31  cudaCtxResetPersistingL2Cache();
32  cuCtxResetPersistingL2Cache();
33  cudaFuncSetAttribute(kernel, cudaFuncAttributeMaxDynamicSharedMemorySize, 8 * 8 * 8);
34  cudaFuncSetAttribute(kernel, cudaFuncAttributePreferredSharedMemoryCarveout, 8 * 8 * 8);
35}

results in the following migrated SYCL code:

 1void bar(int *a) {
 2  /*
 3  DPCT1026:0: The call to __ldg was removed because there is no corresponding
 4  API in SYCL.
 5  */
 6  *a;
 7}
 8
 9void kernel() {
10  return;
11}
12
13void foo() {
14  cudaLimit limit;
15  dpct::queue_ptr stream;
16  unsigned int flags;
17  int config;
18
19  /*
20  DPCT1026:1: The call to cudaDeviceSetLimit was removed because SYCL currently
21  does not support setting resource limits on devices.
22  */
23  /*
24  DPCT1026:2: The call to cudaStreamAttachMemAsync was removed because SYCL
25  currently does not support associating USM with a specific queue.
26  */
27  /*
28  DPCT1026:3: The call to cudaStreamQuery was removed because SYCL currently
29  does not support query operations on queues.
30  */
31  /*
32  DPCT1026:4: The call to cudaDeviceSetSharedMemConfig was removed because SYCL
33  currently does not support configuring shared memory on devices.
34  */
35  /*
36  DPCT1026:5: The call to cudaDeviceSetCacheConfig was removed because SYCL
37  currently does not support setting cache config on devices.
38  */
39  /*
40  DPCT1026:6: The call to cudaSetDeviceFlags was removed because SYCL currently
41  does not support setting flags for devices.
42  */
43  /*
44  DPCT1026:7: The call to cuInit was removed because this call is redundant in
45  SYCL.
46  */
47  float *aptr;
48  /*
49  DPCT1007:8: Migration of cudaStreamAttrValue is not supported.
50  */
51  cudaStreamAttrValue stream_attribute = {};
52  /*
53  DPCT1026:9: The call to cudaStreamSetAttribute was removed because SYCL currently
54  does not support setting cache config on devices.
55  */
56  /*
57  DPCT1026:10: The call to cudaStreamGetAttribute was removed because SYCL currently
58  does not support setting cache config on devices.
59  */
60  /*
61  DPCT1026:11: The call to cudaCtxResetPersistingL2Cache was removed because SYCL
62  currently does not support setting cache config on devices.
63  */
64  /*
65  DPCT1026:12: The call to cuCtxResetPersistingL2Cache was removed because SYCL
66  currently does not support setting cache config on devices.
67  */
68  /*
69  DPCT1026:13: The call to cudaFuncSetAttribute was removed because SYCL currently
70  does not support corresponding setting.
71  */
72  /*
73  DPCT1026:14: The call to cudaFuncSetAttribute was removed because SYCL currently
74  does not support corresponding setting.
75  */
76}

which is rewritten to:

 1void bar(int *a) {
 2  *a;
 3}
 4
 5void foo() {
 6  cudaLimit limit;
 7  dpct::queue_ptr stream;
 8  unsigned int flags;
 9  size_t count;
10  int config;
11  float *aptr;
12
13}