Mikael
I notice that you are using LLVM 3.2. LLVM 3.3 has been released, and I
have been using 3.3 for some time now, and have not recently tested with
3.2. (I'm using OS X as well.)
-erik
On Thu, Jun 27, 2013 at 11:55 AM, Mikael Lepistö
<[email protected]>wrote:
> 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
>
>
--
Erik Schnetter <[email protected]>
http://www.perimeterinstitute.ca/personal/eschnetter/
AIM: eschnett247, Skype: eschnett, Google Talk: [email protected]
------------------------------------------------------------------------------
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
