Skip to content

Commit 79670cb

Browse files
ldorauldorau
committed
Revert "Revert "[UR][SYCL] Introduce UR api to set kernel args + launch in one call." (#19661)"
This reverts commit d25d6d6. Signed-off-by: Lukasz Dorau <[email protected]>
1 parent 1a49fd2 commit 79670cb

File tree

85 files changed

+3383
-435
lines changed

Some content is hidden

Large Commits have some content hidden by default. Use the searchbox below for content that may be hidden.

85 files changed

+3383
-435
lines changed

‎sycl/source/detail/scheduler/commands.cpp‎

Lines changed: 177 additions & 66 deletions
Large diffs are not rendered by default.

‎sycl/source/detail/scheduler/commands.hpp‎

Lines changed: 3 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -633,7 +633,9 @@ void enqueueImpKernel(
633633
ur_kernel_cache_config_t KernelCacheConfig, bool KernelIsCooperative,
634634
const bool KernelUsesClusterLaunch, const size_t WorkGroupMemorySize,
635635
const RTDeviceBinaryImage *BinImage = nullptr,
636-
void *KernelFuncPtr = nullptr);
636+
void *KernelFuncPtr = nullptr, int KernelNumArgs = 0,
637+
detail::kernel_param_desc_t (*KernelParamDescGetter)(int) = nullptr,
638+
bool KernelHasSpecialCaptures = true);
637639

638640
/// The exec CG command enqueues execution of kernel or explicit memory
639641
/// operation.

‎sycl/test-e2e/Adapters/level_zero/batch_barrier.cpp‎

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -24,7 +24,7 @@ int main(int argc, char *argv[]) {
2424
queue q;
2525

2626
submit_kernel(q); // starts a batch
27-
// CHECK: ---> urEnqueueKernelLaunch
27+
// CHECK: ---> urEnqueueKernelLaunchWithArgsExp
2828
// CHECK-NOT: zeCommandQueueExecuteCommandLists
2929

3030
// Initializing Level Zero driver is required if this test is linked
@@ -42,7 +42,7 @@ int main(int argc, char *argv[]) {
4242
// CHECK-NOT: zeCommandQueueExecuteCommandLists
4343

4444
submit_kernel(q);
45-
// CHECK: ---> urEnqueueKernelLaunch
45+
// CHECK: ---> urEnqueueKernelLaunchWithArgsExp
4646
// CHECK-NOT: zeCommandQueueExecuteCommandLists
4747

4848
// interop should close the batch

‎sycl/test-e2e/Adapters/level_zero/batch_test.cpp‎

Lines changed: 32 additions & 33 deletions
Original file line numberDiff line numberDiff line change
@@ -55,55 +55,54 @@
5555
// variable SYCL_PI_LEVEL_ZEOR+BATCH_SIZE=N.
5656
// This test enqueues 8 kernels and then does a wait. And it does this 3 times.
5757
// Expected output is that for batching =1 you will see zeCommandListClose,
58-
// and zeCommandQueueExecuteCommandLists after every urEnqueueKernelLaunch.
59-
// For batching=3 you will see that after 3rd and 6th enqueues, and then after
60-
// urQueueFinish. For 5, after 5th urEnqueue, and then after urQueueFinish. For
61-
// 4 you will see these after 4th and 8th Enqueue, and for 8, only after the
62-
// 8th enqueue. And lastly for 9, you will see the Close and Execute calls
63-
// only after the urQueueFinish.
64-
// Since the test does this 3 times, this pattern will repeat 2 more times,
65-
// and then the test will print Test Passed 8 times, once for each kernel
66-
// validation check.
58+
// and zeCommandQueueExecuteCommandLists after every
59+
// urEnqueueKernelLaunchWithArgsExp. For batching=3 you will see that after 3rd
60+
// and 6th enqueues, and then after urQueueFinish. For 5, after 5th urEnqueue,
61+
// and then after urQueueFinish. For 4 you will see these after 4th and 8th
62+
// Enqueue, and for 8, only after the 8th enqueue. And lastly for 9, you will
63+
// see the Close and Execute calls only after the urQueueFinish. Since the test
64+
// does this 3 times, this pattern will repeat 2 more times, and then the test
65+
// will print Test Passed 8 times, once for each kernel validation check.
6766
// Pattern starts first set of kernel executions.
68-
// CKALL: ---> urEnqueueKernelLaunch
67+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
6968
// CKALL: zeCommandListAppendLaunchKernel(
7069
// CKB1: zeCommandListClose(
7170
// CKB1: zeCommandQueueExecuteCommandLists(
72-
// CKALL: ---> urEnqueueKernelLaunch
71+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
7372
// CKALL: zeCommandListAppendLaunchKernel(
7473
// CKB1: zeCommandListClose(
7574
// CKB1: zeCommandQueueExecuteCommandLists(
76-
// CKALL: ---> urEnqueueKernelLaunch
75+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
7776
// CKALL: zeCommandListAppendLaunchKernel(
7877
// CKB1: zeCommandListClose(
7978
// CKB1: zeCommandQueueExecuteCommandLists(
8079
// CKB3: zeCommandListClose(
8180
// CKB3: zeCommandQueueExecuteCommandLists(
82-
// CKALL: ---> urEnqueueKernelLaunch
81+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
8382
// CKALL: zeCommandListAppendLaunchKernel(
8483
// CKB1: zeCommandListClose(
8584
// CKB1: zeCommandQueueExecuteCommandLists(
8685
// CKB4: zeCommandListClose(
8786
// CKB4: zeCommandQueueExecuteCommandLists(
88-
// CKALL: ---> urEnqueueKernelLaunch
87+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
8988
// CKALL: zeCommandListAppendLaunchKernel(
9089
// CKB1: zeCommandListClose(
9190
// CKB1: zeCommandQueueExecuteCommandLists(
9291
// CKB5: zeCommandListClose(
9392
// CKB5: zeCommandQueueExecuteCommandLists(
94-
// CKALL: ---> urEnqueueKernelLaunch
93+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
9594
// CKALL: zeCommandListAppendLaunchKernel(
9695
// CKB1: zeCommandListClose(
9796
// CKB1: zeCommandQueueExecuteCommandLists(
9897
// CKB3: zeCommandListClose(
9998
// CKB3: zeCommandQueueExecuteCommandLists(
100-
// CKALL: ---> urEnqueueKernelLaunch
99+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
101100
// CKALL: zeCommandListAppendLaunchKernel(
102101
// CKB1: zeCommandListClose(
103102
// CKB1: zeCommandQueueExecuteCommandLists(
104103
// CKB7: zeCommandListClose(
105104
// CKB7: zeCommandQueueExecuteCommandLists(
106-
// CKALL: ---> urEnqueueKernelLaunch
105+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
107106
// CKALL: zeCommandListAppendLaunchKernel(
108107
// CKB1: zeCommandListClose(
109108
// CKB1: zeCommandQueueExecuteCommandLists(
@@ -121,45 +120,45 @@
121120
// CKB9: zeCommandListClose(
122121
// CKB9: zeCommandQueueExecuteCommandLists(
123122
// Pattern starts 2nd set of kernel executions
124-
// CKALL: ---> urEnqueueKernelLaunch
123+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
125124
// CKALL: zeCommandListAppendLaunchKernel(
126125
// CKB1: zeCommandListClose(
127126
// CKB1: zeCommandQueueExecuteCommandLists(
128-
// CKALL: ---> urEnqueueKernelLaunch
127+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
129128
// CKALL: zeCommandListAppendLaunchKernel(
130129
// CKB1: zeCommandListClose(
131130
// CKB1: zeCommandQueueExecuteCommandLists(
132-
// CKALL: ---> urEnqueueKernelLaunch
131+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
133132
// CKALL: zeCommandListAppendLaunchKernel(
134133
// CKB1: zeCommandListClose(
135134
// CKB1: zeCommandQueueExecuteCommandLists(
136135
// CKB3: zeCommandListClose(
137136
// CKB3: zeCommandQueueExecuteCommandLists(
138-
// CKALL: ---> urEnqueueKernelLaunch
137+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
139138
// CKALL: zeCommandListAppendLaunchKernel(
140139
// CKB1: zeCommandListClose(
141140
// CKB1: zeCommandQueueExecuteCommandLists(
142141
// CKB4: zeCommandListClose(
143142
// CKB4: zeCommandQueueExecuteCommandLists(
144-
// CKALL: ---> urEnqueueKernelLaunch
143+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
145144
// CKALL: zeCommandListAppendLaunchKernel(
146145
// CKB1: zeCommandListClose(
147146
// CKB1: zeCommandQueueExecuteCommandLists(
148147
// CKB5: zeCommandListClose(
149148
// CKB5: zeCommandQueueExecuteCommandLists(
150-
// CKALL: ---> urEnqueueKernelLaunch
149+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
151150
// CKALL: zeCommandListAppendLaunchKernel(
152151
// CKB1: zeCommandListClose(
153152
// CKB1: zeCommandQueueExecuteCommandLists(
154153
// CKB3: zeCommandListClose(
155154
// CKB3: zeCommandQueueExecuteCommandLists(
156-
// CKALL: ---> urEnqueueKernelLaunch
155+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
157156
// CKALL: zeCommandListAppendLaunchKernel(
158157
// CKB1: zeCommandListClose(
159158
// CKB1: zeCommandQueueExecuteCommandLists(
160159
// CKB7: zeCommandListClose(
161160
// CKB7: zeCommandQueueExecuteCommandLists(
162-
// CKALL: ---> urEnqueueKernelLaunch
161+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
163162
// CKALL: zeCommandListAppendLaunchKernel(
164163
// CKB1: zeCommandListClose(
165164
// CKB1: zeCommandQueueExecuteCommandLists(
@@ -177,45 +176,45 @@
177176
// CKB9: zeCommandListClose(
178177
// CKB9: zeCommandQueueExecuteCommandLists(
179178
// Pattern starts 3rd set of kernel executions
180-
// CKALL: ---> urEnqueueKernelLaunch
179+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
181180
// CKALL: zeCommandListAppendLaunchKernel(
182181
// CKB1: zeCommandListClose(
183182
// CKB1: zeCommandQueueExecuteCommandLists(
184-
// CKALL: ---> urEnqueueKernelLaunch
183+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
185184
// CKALL: zeCommandListAppendLaunchKernel(
186185
// CKB1: zeCommandListClose(
187186
// CKB1: zeCommandQueueExecuteCommandLists(
188-
// CKALL: ---> urEnqueueKernelLaunch
187+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
189188
// CKALL: zeCommandListAppendLaunchKernel(
190189
// CKB1: zeCommandListClose(
191190
// CKB1: zeCommandQueueExecuteCommandLists(
192191
// CKB3: zeCommandListClose(
193192
// CKB3: zeCommandQueueExecuteCommandLists(
194-
// CKALL: ---> urEnqueueKernelLaunch
193+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
195194
// CKALL: zeCommandListAppendLaunchKernel(
196195
// CKB1: zeCommandListClose(
197196
// CKB1: zeCommandQueueExecuteCommandLists(
198197
// CKB4: zeCommandListClose(
199198
// CKB4: zeCommandQueueExecuteCommandLists(
200-
// CKALL: ---> urEnqueueKernelLaunch
199+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
201200
// CKALL: zeCommandListAppendLaunchKernel(
202201
// CKB1: zeCommandListClose(
203202
// CKB1: zeCommandQueueExecuteCommandLists(
204203
// CKB5: zeCommandListClose(
205204
// CKB5: zeCommandQueueExecuteCommandLists(
206-
// CKALL: ---> urEnqueueKernelLaunch
205+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
207206
// CKALL: zeCommandListAppendLaunchKernel(
208207
// CKB1: zeCommandListClose(
209208
// CKB1: zeCommandQueueExecuteCommandLists(
210209
// CKB3: zeCommandListClose(
211210
// CKB3: zeCommandQueueExecuteCommandLists(
212-
// CKALL: ---> urEnqueueKernelLaunch
211+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
213212
// CKALL: zeCommandListAppendLaunchKernel(
214213
// CKB1: zeCommandListClose(
215214
// CKB1: zeCommandQueueExecuteCommandLists(
216215
// CKB7: zeCommandListClose(
217216
// CKB7: zeCommandQueueExecuteCommandLists(
218-
// CKALL: ---> urEnqueueKernelLaunch
217+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
219218
// CKALL: zeCommandListAppendLaunchKernel(
220219
// CKB1: zeCommandListClose(
221220
// CKB1: zeCommandQueueExecuteCommandLists(

‎sycl/test-e2e/Adapters/level_zero/batch_test_copy_with_compute.cpp‎

Lines changed: 32 additions & 34 deletions
Original file line numberDiff line numberDiff line change
@@ -31,55 +31,53 @@
3131
// variable SYCL_PI_LEVEL_ZERO_{COPY_}BATCH_SIZE=N.
3232
// This test enqueues 8 kernels and then does a wait. And it does this 3 times.
3333
// Expected output is that for batching =1 you will see zeCommandListClose,
34-
// and zeCommandQueueExecuteCommandLists after every urEnqueueKernelLaunch.
35-
// For batching=3 you will see that after 3rd and 6th enqueues, and then after
36-
// urEventWait. For 5, after 5th urEnqueue, and then after urEventWait. For
37-
// 4 you will see these after 4th and 8th Enqueue, and for 8, only after the
38-
// 8th enqueue. And lastly for 9, you will see the Close and Execute calls
39-
// only after the urEventWait.
40-
// Since the test does this 3 times, this pattern will repeat 2 more times,
41-
// and then the test will print Test Passed 8 times, once for each kernel
42-
// validation check.
43-
// Pattern starts first set of kernel executions.
44-
// CKALL: ---> urEnqueueKernelLaunch
34+
// and zeCommandQueueExecuteCommandLists after every
35+
// urEnqueueKernelLaunchWithArgsExp. For batching=3 you will see that after 3rd
36+
// and 6th enqueues, and then after urEventWait. For 5, after 5th urEnqueue, and
37+
// then after urEventWait. For 4 you will see these after 4th and 8th Enqueue,
38+
// and for 8, only after the 8th enqueue. And lastly for 9, you will see the
39+
// Close and Execute calls only after the urEventWait. Since the test does this
40+
// 3 times, this pattern will repeat 2 more times, and then the test will print
41+
// Test Passed 8 times, once for each kernel validation check. Pattern starts
42+
// first set of kernel executions. CKALL: ---> urEnqueueKernelLaunchWithArgsExp
4543
// CKALL: zeCommandListAppendLaunchKernel(
4644
// CKB1: zeCommandListClose(
4745
// CKB1: zeCommandQueueExecuteCommandLists(
48-
// CKALL: ---> urEnqueueKernelLaunch
46+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
4947
// CKALL: zeCommandListAppendLaunchKernel(
5048
// CKB1: zeCommandListClose(
5149
// CKB1: zeCommandQueueExecuteCommandLists(
52-
// CKALL: ---> urEnqueueKernelLaunch
50+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
5351
// CKALL: zeCommandListAppendLaunchKernel(
5452
// CKB1: zeCommandListClose(
5553
// CKB1: zeCommandQueueExecuteCommandLists(
5654
// CKB3: zeCommandListClose(
5755
// CKB3: zeCommandQueueExecuteCommandLists(
58-
// CKALL: ---> urEnqueueKernelLaunch
56+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
5957
// CKALL: zeCommandListAppendLaunchKernel(
6058
// CKB1: zeCommandListClose(
6159
// CKB1: zeCommandQueueExecuteCommandLists(
6260
// CKB4: zeCommandListClose(
6361
// CKB4: zeCommandQueueExecuteCommandLists(
64-
// CKALL: ---> urEnqueueKernelLaunch
62+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
6563
// CKALL: zeCommandListAppendLaunchKernel(
6664
// CKB1: zeCommandListClose(
6765
// CKB1: zeCommandQueueExecuteCommandLists(
6866
// CKB5: zeCommandListClose(
6967
// CKB5: zeCommandQueueExecuteCommandLists(
70-
// CKALL: ---> urEnqueueKernelLaunch
68+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
7169
// CKALL: zeCommandListAppendLaunchKernel(
7270
// CKB1: zeCommandListClose(
7371
// CKB1: zeCommandQueueExecuteCommandLists(
7472
// CKB3: zeCommandListClose(
7573
// CKB3: zeCommandQueueExecuteCommandLists(
76-
// CKALL: ---> urEnqueueKernelLaunch
74+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
7775
// CKALL: zeCommandListAppendLaunchKernel(
7876
// CKB1: zeCommandListClose(
7977
// CKB1: zeCommandQueueExecuteCommandLists(
8078
// CKB7: zeCommandListClose(
8179
// CKB7: zeCommandQueueExecuteCommandLists(
82-
// CKALL: ---> urEnqueueKernelLaunch
80+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
8381
// CKALL: zeCommandListAppendLaunchKernel(
8482
// CKB1: zeCommandListClose(
8583
// CKB1: zeCommandQueueExecuteCommandLists(
@@ -97,45 +95,45 @@
9795
// CKB9: zeCommandListClose(
9896
// CKB9: zeCommandQueueExecuteCommandLists(
9997
// Pattern starts 2nd set of kernel executions
100-
// CKALL: ---> urEnqueueKernelLaunch
98+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
10199
// CKALL: zeCommandListAppendLaunchKernel(
102100
// CKB1: zeCommandListClose(
103101
// CKB1: zeCommandQueueExecuteCommandLists(
104-
// CKALL: ---> urEnqueueKernelLaunch
102+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
105103
// CKALL: zeCommandListAppendLaunchKernel(
106104
// CKB1: zeCommandListClose(
107105
// CKB1: zeCommandQueueExecuteCommandLists(
108-
// CKALL: ---> urEnqueueKernelLaunch
106+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
109107
// CKALL: zeCommandListAppendLaunchKernel(
110108
// CKB1: zeCommandListClose(
111109
// CKB1: zeCommandQueueExecuteCommandLists(
112110
// CKB3: zeCommandListClose(
113111
// CKB3: zeCommandQueueExecuteCommandLists(
114-
// CKALL: ---> urEnqueueKernelLaunch
112+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
115113
// CKALL: zeCommandListAppendLaunchKernel(
116114
// CKB1: zeCommandListClose(
117115
// CKB1: zeCommandQueueExecuteCommandLists(
118116
// CKB4: zeCommandListClose(
119117
// CKB4: zeCommandQueueExecuteCommandLists(
120-
// CKALL: ---> urEnqueueKernelLaunch
118+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
121119
// CKALL: zeCommandListAppendLaunchKernel(
122120
// CKB1: zeCommandListClose(
123121
// CKB1: zeCommandQueueExecuteCommandLists(
124122
// CKB5: zeCommandListClose(
125123
// CKB5: zeCommandQueueExecuteCommandLists(
126-
// CKALL: ---> urEnqueueKernelLaunch
124+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
127125
// CKALL: zeCommandListAppendLaunchKernel(
128126
// CKB1: zeCommandListClose(
129127
// CKB1: zeCommandQueueExecuteCommandLists(
130128
// CKB3: zeCommandListClose(
131129
// CKB3: zeCommandQueueExecuteCommandLists(
132-
// CKALL: ---> urEnqueueKernelLaunch
130+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
133131
// CKALL: zeCommandListAppendLaunchKernel(
134132
// CKB1: zeCommandListClose(
135133
// CKB1: zeCommandQueueExecuteCommandLists(
136134
// CKB7: zeCommandListClose(
137135
// CKB7: zeCommandQueueExecuteCommandLists(
138-
// CKALL: ---> urEnqueueKernelLaunch
136+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
139137
// CKALL: zeCommandListAppendLaunchKernel(
140138
// CKB1: zeCommandListClose(
141139
// CKB1: zeCommandQueueExecuteCommandLists(
@@ -153,45 +151,45 @@
153151
// CKB9: zeCommandListClose(
154152
// CKB9: zeCommandQueueExecuteCommandLists(
155153
// Pattern starts 3rd set of kernel executions
156-
// CKALL: ---> urEnqueueKernelLaunch
154+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
157155
// CKALL: zeCommandListAppendLaunchKernel(
158156
// CKB1: zeCommandListClose(
159157
// CKB1: zeCommandQueueExecuteCommandLists(
160-
// CKALL: ---> urEnqueueKernelLaunch
158+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
161159
// CKALL: zeCommandListAppendLaunchKernel(
162160
// CKB1: zeCommandListClose(
163161
// CKB1: zeCommandQueueExecuteCommandLists(
164-
// CKALL: ---> urEnqueueKernelLaunch
162+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
165163
// CKALL: zeCommandListAppendLaunchKernel(
166164
// CKB1: zeCommandListClose(
167165
// CKB1: zeCommandQueueExecuteCommandLists(
168166
// CKB3: zeCommandListClose(
169167
// CKB3: zeCommandQueueExecuteCommandLists(
170-
// CKALL: ---> urEnqueueKernelLaunch
168+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
171169
// CKALL: zeCommandListAppendLaunchKernel(
172170
// CKB1: zeCommandListClose(
173171
// CKB1: zeCommandQueueExecuteCommandLists(
174172
// CKB4: zeCommandListClose(
175173
// CKB4: zeCommandQueueExecuteCommandLists(
176-
// CKALL: ---> urEnqueueKernelLaunch
174+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
177175
// CKALL: zeCommandListAppendLaunchKernel(
178176
// CKB1: zeCommandListClose(
179177
// CKB1: zeCommandQueueExecuteCommandLists(
180178
// CKB5: zeCommandListClose(
181179
// CKB5: zeCommandQueueExecuteCommandLists(
182-
// CKALL: ---> urEnqueueKernelLaunch
180+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
183181
// CKALL: zeCommandListAppendLaunchKernel(
184182
// CKB1: zeCommandListClose(
185183
// CKB1: zeCommandQueueExecuteCommandLists(
186184
// CKB3: zeCommandListClose(
187185
// CKB3: zeCommandQueueExecuteCommandLists(
188-
// CKALL: ---> urEnqueueKernelLaunch
186+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
189187
// CKALL: zeCommandListAppendLaunchKernel(
190188
// CKB1: zeCommandListClose(
191189
// CKB1: zeCommandQueueExecuteCommandLists(
192190
// CKB7: zeCommandListClose(
193191
// CKB7: zeCommandQueueExecuteCommandLists(
194-
// CKALL: ---> urEnqueueKernelLaunch
192+
// CKALL: ---> urEnqueueKernelLaunchWithArgsExp
195193
// CKALL: zeCommandListAppendLaunchKernel(
196194
// CKB1: zeCommandListClose(
197195
// CKB1: zeCommandQueueExecuteCommandLists(

0 commit comments

Comments
 (0)