(original) (raw)

Hi Justin,


Thanks so much for the response. I'll try that out and get back to you.

Sincerely,
Michael Ehmke

On Wed, Nov 29, 2017, 12:06 AM Justin Lebar <jlebar@google.com> wrote:
Hi, Michael.

Are you able to create a reduced testcase which demonstrates your problem?

I tried

template
__global__ void Kernel() {}

void test() {
Kernel<16><<<100, 200>>>();
}

which I compiled with

$ clang -c -Xclang -ast-dump test.cu

and the AST looks fine to me -- the CUDAKernelCallExpr shows up as expected.

-Justin

On Tue, Nov 28, 2017 at 2:33 PM Michael Ehmke via llvm-dev <llvm-dev@lists.llvm.org> wrote:
Hello all,

I'm using Clang and am trying to refactor CUDA code. I want to traverse the AST to access a CUDAKernelCallExpr node where the kernel call has template parameters. I have been able to successfully match on and access CUDA kernel calls which have no template parameters by using the AST matchers to match on a CUDAKernelCallExpr. However, when my code comes across a kernel call with template parameters, it appears that the expression is not even existent in the AST when it is dumped.

This is the expression that I'm having trouble parsing:

ex. matrixMulCUDA<16><<< grid, threads >>>(d_C, d_A, d_B, dimsA.x, dimsB.x);

This expression does not show up at all in the AST when dumping the AST. I have been able to successfully parse the following kernel call expression and it shows up in the AST as a "CudaKernelCallExpr" node.

ex. matrixMulCUDA<<< grid, threads >>>(d_C, d_A, d_B, dimsA.x, dimsB.x);

As you can see, the only difference is the <16> after the name of the call, and this node is clearly represented in the AST.

Here is a screenshot showing the lack of a node to represent the call with template parameters.

image.png
If I could get any assistance on why this is happening, I would appreciate it greatly.

Thank you,
Michael Ehmke

_______________________________________________

LLVM Developers mailing list

llvm-dev@lists.llvm.org

http://lists.llvm.org/cgi-bin/mailman/listinfo/llvm-dev