Hi,
I was able to limit the case where I got pocl-bzr to crash on osx. I woudl
appreciate if someone could reproduce the error.
Thanks, Mikael
#define TABLE_SIZE 20
typedef struct {
uint table[TABLE_SIZE];
} WclPrivates;
typedef struct {
uint dummy;
} WclConstants;
typedef struct {
__constant WclConstants *constantAllocations_min;
__constant WclConstants *constantAllocations_max;
} WclConstantLimits;
typedef struct {
WclConstantLimits constantLimits;
WclPrivates privateAllocations;
} WclProgramAllocations;
__constant WclConstants constantAllocations = { 0 };
__kernel
void test_address_space_structs(__global uint *out, int num_workgroups)
{
WclProgramAllocations wcl_limits;
// setup constant allocation areas
wcl_limits.constantLimits.constantAllocations_min =
&constantAllocations;
wcl_limits.constantLimits.constantAllocations_max =
(&constantAllocations) + 1;
const size_t work_item_base = get_local_id(0);
// Copy work item result into work group result.
for (size_t i = 0; i < TABLE_SIZE; ++i) {
const size_t index = work_item_base + i;
out[index] =
wcl_limits.privateAllocations.table[i];
}
}
------ error -------
Running test #0 test_address_space_structs...
Assertion failed: (getOperand(0)->getType() ==
cast<PointerType>(getOperand(1)->getType())->getElementType() && "Ptr must
be a pointer to Val type!"), function AssertOK, file
/Users/mikaelle/Projects/Vincit/webcl/llvm-3.2-final/lib/VMCore/Instructions.cpp,
line 1080.
0 libLLVM-3.2svn.dylib 0x00000001013dc8de PrintStackTrace(void*) + 46
1 libLLVM-3.2svn.dylib 0x00000001013dce89 SignalHandler(int) + 297
2 libsystem_c.dylib 0x00007fff8f57294a _sigtramp + 26
3 libsystem_c.dylib 0x00007fff5f83d248 _sigtramp + 18446744072907172120
4 libLLVM-3.2svn.dylib 0x00000001013dcbab raise + 27
5 libLLVM-3.2svn.dylib 0x00000001013dcc62 abort + 18
6 libLLVM-3.2svn.dylib 0x00000001013dcc41 __assert_rtn + 129
7 libLLVM-3.2svn.dylib 0x0000000100d4f37f llvm::StoreInst::AssertOK() + 415
8 libLLVM-3.2svn.dylib 0x0000000100d4fc89
llvm::StoreInst::StoreInst(llvm::Value*, llvm::Value*, bool, unsigned int,
llvm::AtomicOrdering, llvm::SynchronizationScope, llvm::Instruction*) + 281
9 libLLVM-3.2svn.dylib 0x0000000100d4fb5d
llvm::StoreInst::StoreInst(llvm::Value*, llvm::Value*, bool, unsigned int,
llvm::AtomicOrdering, llvm::SynchronizationScope, llvm::Instruction*) + 109
10 libLLVM-3.2svn.dylib 0x0000000100d5b98c llvm::StoreInst::clone_impl()
const + 172
11 libLLVM-3.2svn.dylib 0x0000000100d4893f llvm::Instruction::clone() const
+ 63
12 libLLVM-3.2svn.dylib 0x0000000101430042
llvm::CloneBasicBlock(llvm::BasicBlock const*, llvm::ValueMap<llvm::Value
const*, llvm::WeakVH, llvm::ValueMapConfig<llvm::Value const*> >&,
llvm::Twine const&, llvm::Function*, llvm::ClonedCodeInfo*) + 338
13 libLLVM-3.2svn.dylib 0x000000010143074e
llvm::CloneFunctionInto(llvm::Function*, llvm::Function const*,
llvm::ValueMap<llvm::Value const*, llvm::WeakVH,
llvm::ValueMapConfig<llvm::Value const*> >&, bool,
llvm::SmallVectorImpl<llvm::ReturnInst*>&, char const*,
llvm::ClonedCodeInfo*, llvm::ValueMapTypeRemapper*) + 1086
14 llvmopencl.so 0x00000001039f2370
pocl::TargetAddressSpaces::runOnModule(llvm::Module&) + 1926
15 libLLVM-3.2svn.dylib 0x0000000100d8596d
llvm::MPPassManager::runOnModule(llvm::Module&) + 493
16 libLLVM-3.2svn.dylib 0x0000000100d860a7
llvm::PassManagerImpl::run(llvm::Module&) + 167
17 libLLVM-3.2svn.dylib 0x0000000100d862c1
llvm::PassManager::run(llvm::Module&) + 33
18 opt 0x00000001003d3f7c main + 6588
19 libdyld.dylib 0x00007fff8a3727e1 start + 0
20 libdyld.dylib 0x0000000000000025 start + 18446603338197293124
Stack dump:
0. Program arguments:
/Users/mikaelle/Projects/Vincit/webcl/install-3.2/bin/opt
-load=/Users/mikaelle/Projects/Vincit/webcl/pocl-install/lib/pocl/llvmopencl.so
-domtree -workitem-handler-chooser -break-constgeps -generate-header
-flatten -always-inline -globaldce -simplifycfg -loop-simplify
-phistoallocas -isolate-regions -uniformity -implicit-loop-barriers
-loop-barriers -barriertails -barriers -isolate-regions -add-wi-metadata
-wi-aa -workitemrepl -workitemloops -allocastoentry -workgroup
-kernel=test_address_space_structs -local-size=1 1 1
-disable-simplify-libcalls -target-address-spaces -O3 -instcombine
-header=/tmp/poclLhuE10/pthread/test_address_space_structs/kernel_header.h
-o
/tmp/poclLhuE10/pthread/test_address_space_structs/1-1-1.0-0-0/parallel.bc
/tmp/poclLhuE10/pthread/test_address_space_structs/kernel_linked.bc
1. Running pass 'Convert the 'fake' address space ids to the target
specific ones.' on module
'/tmp/poclLhuE10/pthread/test_address_space_structs/kernel_linked.bc'.
/Users/mikaelle/Projects/Vincit/webcl/pocl-install/share/pocl/pocl-workgroup:
line 143: 77490 Illegal instruction: 4
/Users/mikaelle/Projects/Vincit/webcl/install-3.2/bin/opt
-load=${pocl_lib} -domtree -workitem-handler-chooser -break-constgeps
-generate-header -flatten -always-inline -globaldce -simplifycfg
-loop-simplify -phistoallocas -isolate-regions -uniformity
-implicit-loop-barriers -loop-barriers -barriertails -barriers
-isolate-regions -add-wi-metadata -wi-aa -workitemrepl -workitemloops
-allocastoentry -workgroup -kernel=${kernel} -local-size=${size_x}
${size_y} ${size_z} -disable-simplify-libcalls -target-address-spaces
${OPT_SWITCH} ${WG_VECTORIZER} ${EXTRA_OPTS} -instcombine -header=${header}
-o ${output_file} ${linked_bc}
clEnqueueNDRangeKernel call failed
------------------------------------------------------------------------------
This SF.net email is sponsored by Windows:
Build for Windows Store.
http://p.sf.net/sfu/windows-dev2dev
_______________________________________________
pocl-devel mailing list
[email protected]
https://lists.sourceforge.net/lists/listinfo/pocl-devel
