НЕДОСТУПНО выполнено! ошибка при попытке создать PTX

Question

Я пытаюсь сгенерировать код PTX для ядра примера программы "nbody" (nbody_kernel.cu) с использованием clang/LLVM версии 3.2. Программа nbody CUDA доступна в SDK от Nvidia.

Я имею в виду проект https://github.com/jholewinski/llvm-ptx-samples.

Ниже приведены мои команды,

clang ++ -O4 -S -I / usr / local / cuda / include -emit-llvm -target nvptx64 nbody_kernel.cu -o nbody_kernel.ll

opt -O3 -loop-unroll -unroll-allow-частичное nbody_kernel.ll -o nbody_kernel.ll

llc nbody_kernel.ll -o nbody_kernel.ptx

После выполнения последней команды (ООО) я получаю НЕПРЕРЫВНОЕ выполнение! ошибка со следующей трассировкой стека

UNREACHABLE executed!
0  libLLVM-3.2svn.so 0x00000033eb79fadf
1  libLLVM-3.2svn.so 0x00000033eb7a01d6
2  libpthread.so.0   0x0000003825a0f500
3  libc.so.6         0x0000003824e328a5 gsignal + 53
4  libc.so.6         0x0000003824e34085 abort + 373
5  libLLVM-3.2svn.so 0x00000033eb788320 llvm::llvm_unreachable_internal(char const*, char const*, unsigned int) + 336
6  libLLVM-3.2svn.so 0x00000033eb43dc37
7  libLLVM-3.2svn.so 0x00000033eb4411a0
8  libLLVM-3.2svn.so 0x00000033eb4422a5
9  libLLVM-3.2svn.so 0x00000033eb210d76 llvm::FPPassManager::doInitialization(llvm::Module&) + 54
10 libLLVM-3.2svn.so 0x00000033eb217ec7 llvm::FPPassManager::runOnModule(llvm::Module&) + 23
11 libLLVM-3.2svn.so 0x00000033eb217ac5 llvm::MPPassManager::runOnModule(llvm::Module&) + 389
12 libLLVM-3.2svn.so 0x00000033eb217c1b llvm::PassManagerImpl::run(llvm::Module&) + 107
13 llc               0x000000000040d157 main + 5415
14 libc.so.6         0x0000003824e1ecdd __libc_start_main + 253
15 llc               0x0000000000409a29
Stack dump:
0.  Program arguments: llc nbody_kernel.ll -o nbody_kernel.ptx 
1.  Running pass 'Function Pass Manager' on module 'nbody_kernel.ll'.
Aborted (core dumped)

Я заменил глобальные индексы, например, для threadIdx.x, на __builtin_ptx_read_tid_x() и другие. Нет проблем с генерацией IR LLVM (т.е..ll). Ошибка появляется при попытке сгенерировать PTX из IR с помощью llc.

Есть какие-нибудь указатели на то, что может происходить здесь?