https://bugs.llvm.org/show_bug.cgi?id=48866
Bug ID: 48866
Summary: Kernel launch failure with lambdas in CUDA on Windows
Product: clang
Version: unspecified
Hardware: PC
OS: Windows NT
Status: NEW
Severity: normal
Priority: P
Component: CUDA
Assignee: unassignedclangb...@nondot.org
Reporter: joac...@joameyer.de
CC: llvm-bugs@lists.llvm.org
Given multiple (nested) lambdas as CUDA kernel parameters, the mangling/naming
seems to be inconsistent between host and device and thus `cudaLaunchKernel`
fails with error code 98. Might be related to the lambda numbering issues,
solved in Clang 10 (https://reviews.llvm.org/D68818)?
A rather minimal reproducer is:
```
#include <exception>
#include <iostream>
template <class T>
__global__ void kernel(int *t, int n, T lambda)
{
if(threadIdx.x < n)
t[threadIdx.x] = lambda(t[threadIdx.x]);
}
int main(){
int* t=nullptr;
cudaMalloc(&t,12);
auto lambda = [](int t){
return t++;
};
kernel<<<1, 32>>>(t, 3, lambda);
[&]()
{
auto lambda2 = [](int t){
return t++;
};
kernel<<<1, 32>>>(t, 3, lambda2);
}();
if(cudaGetLastError())
{
std::cout << "err";
cudaFree(t);
std::terminate();
}
cudaFree(t);
return 0;
}
```
--
You are receiving this mail because:
You are on the CC list for the bug._______________________________________________
llvm-bugs mailing list
llvm-bugs@lists.llvm.org
https://lists.llvm.org/cgi-bin/mailman/listinfo/llvm-bugs
