Skip to content

Commit a92ff30

Browse files
committed
Store hasSpecialCaptures in kernel information
We are currently only able to skip argument extraction in the case where a lambda has no special captures. We can only detect this while we have the kernel type name, and must carry it through until we call handler::finalize(). Signed-off-by: John Pennycook <[email protected]>
1 parent d046b0a commit a92ff30

File tree

3 files changed

+6
-3
lines changed

3 files changed

+6
-3
lines changed

sycl/include/sycl/handler.hpp

Lines changed: 3 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -799,7 +799,8 @@ class __SYCL_EXPORT handler {
799799
setKernelInfo((void *)MHostKernel->getPtr(),
800800
detail::getKernelNumParams<KernelName>(),
801801
&(detail::getKernelParamDesc<KernelName>),
802-
detail::isKernelESIMD<KernelName>());
802+
detail::isKernelESIMD<KernelName>(),
803+
detail::hasSpecialCaptures<KernelName>());
803804

804805
MKernelName = detail::getKernelName<KernelName>();
805806
} else {
@@ -3763,7 +3764,7 @@ class __SYCL_EXPORT handler {
37633764

37643765
void setKernelInfo(void *KernelFuncPtr, int KernelNumArgs,
37653766
detail::kernel_param_desc_t (*KernelParamDescGetter)(int),
3766-
bool KernelIsESIMD);
3767+
bool KernelIsESIMD, bool KernelHasSpecialCaptures);
37673768

37683769
friend class detail::HandlerAccess;
37693770

sycl/source/detail/handler_impl.hpp

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -204,6 +204,7 @@ class handler_impl {
204204
int MKernelNumArgs = 0;
205205
detail::kernel_param_desc_t (*MKernelParamDescGetter)(int) = nullptr;
206206
bool MKernelIsESIMD = false;
207+
bool MKernelHasSpecialCaptures = false;
207208
};
208209

209210
} // namespace detail

sycl/source/handler.cpp

Lines changed: 2 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -2266,11 +2266,12 @@ void handler::setNDRangeDescriptorPadded(sycl::range<3> NumWorkItems,
22662266
void handler::setKernelInfo(
22672267
void *KernelFuncPtr, int KernelNumArgs,
22682268
detail::kernel_param_desc_t (*KernelParamDescGetter)(int),
2269-
bool KernelIsESIMD) {
2269+
bool KernelIsESIMD, bool KernelHasSpecialCaptures) {
22702270
impl->MKernelFuncPtr = KernelFuncPtr;
22712271
impl->MKernelNumArgs = KernelNumArgs;
22722272
impl->MKernelParamDescGetter = KernelParamDescGetter;
22732273
impl->MKernelIsESIMD = KernelIsESIMD;
2274+
impl->MKernelHasSpecialCaptures = KernelHasSpecialCaptures;
22742275
}
22752276

22762277
void handler::saveCodeLoc(detail::code_location CodeLoc, bool IsTopCodeLoc) {

0 commit comments

Comments
 (0)