<html>
<head>
<base href="https://llvm.org/bugs/" />
</head>
<body><table border="1" cellspacing="0" cellpadding="8">
<tr>
<th>Bug ID</th>
<td><a class="bz_bug_link
bz_status_NEW "
title="NEW --- - Cuda __device__ lambda does not work"
href="https://llvm.org/bugs/show_bug.cgi?id=26341">26341</a>
</td>
</tr>
<tr>
<th>Summary</th>
<td>Cuda __device__ lambda does not work
</td>
</tr>
<tr>
<th>Product</th>
<td>clang
</td>
</tr>
<tr>
<th>Version</th>
<td>trunk
</td>
</tr>
<tr>
<th>Hardware</th>
<td>PC
</td>
</tr>
<tr>
<th>OS</th>
<td>Linux
</td>
</tr>
<tr>
<th>Status</th>
<td>NEW
</td>
</tr>
<tr>
<th>Severity</th>
<td>normal
</td>
</tr>
<tr>
<th>Priority</th>
<td>P
</td>
</tr>
<tr>
<th>Component</th>
<td>-New Bugs
</td>
</tr>
<tr>
<th>Assignee</th>
<td>unassignedclangbugs@nondot.org
</td>
</tr>
<tr>
<th>Reporter</th>
<td>crtrott@sandia.gov
</td>
</tr>
<tr>
<th>CC</th>
<td>llvm-bugs@lists.llvm.org
</td>
</tr>
<tr>
<th>Classification</th>
<td>Unclassified
</td>
</tr></table>
<p>
<div>
<pre>Created <span class=""><a href="attachment.cgi?id=15731" name="attach_15731" title="Reproduction code">attachment 15731</a> <a href="attachment.cgi?id=15731&action=edit" title="Reproduction code">[details]</a></span>
Reproduction code
This is an experimental feature in CUDA 7.5 and expected to be non experimental
in CUDA 8. The attached simple test code fails to compile.
Code:
run( [=] __device__ (int i) {
d_a[i] = d_c;
});
Error:
main.cpp:23:12: error: lambda requires '()' before attribute specifier
run( [=] __device__ (int i) {
^
()
/usr/local/cuda/include/host_defines.h:189:9: note: expanded from macro
'__device__'
__location__(device)
^
/usr/local/cuda/include/host_defines.h:88:9: note: expanded from macro
'__location__'
__annotate__(a)
^
/usr/local/cuda/include/host_defines.h:86:9: note: expanded from macro
'__annotate__'
__attribute__((a))
^
main.cpp:23:23: error: expected body of lambda expression
run( [=] __device__ (int i) {
^
main.cpp:23:28: error: expected '(' for function-style cast or type
construction
run( [=] __device__ (int i) {
~~~ ^
3 errors generated.
I actually like to see to be able to do __host__ __device__ (which does not
work with NVCC right now) so that you can dispatch the lambda to the host or to
the device. Effectively I just want the operator of the auto generated struct
to be marked __host__ __device__. I am responsible to only capture stuff which
is accesible on the device. Furthermore one might want to restrict a __device__
lambda to only capture by value. i.e. [=] __host__ __device__ (...) {...}</pre>
</div>
</p>
<hr>
<span>You are receiving this mail because:</span>
<ul>
<li>You are on the CC list for the bug.</li>
</ul>
</body>
</html>