maintainer-feedback requested: [Bug 219561] [lang/clover] partially broken

[bugzilla-noreply at freebsd.org]

Johannes M Dieterich <jmd at freebsd.org> has reassigned Bugzilla Automation
<bugzilla at FreeBSD.org>'s request for maintainer-feedback to x11 at FreeBSD.org:
Bug 219561: [lang/clover] partially broken
https://bugs.freebsd.org/bugzilla/show_bug.cgi?id=219561



--- Description ---
On FreeBSDDesktop version of HEAD, lang/clover does not do double precision for
at least Carrizo properly:

LLVM ERROR: Cannot select: t708: f64 = fpow t680, t771
  t680: f64 = bitcast t119
    t119: i64 = build_pair t880, t881
      t880: i32 = extract_vector_elt t842, Constant:i32<0>
	t842: v4i32,ch = load<LD16[%arrayidx(addrspace=1)+64](align=64)> t112,
undef:i64
	  t62: i64 = add t38, Constant:i64<64>
	    t38: i64 = add t773, t37
	      t773: i64 = bitcast t772
		t772: v2i32,ch =
load<LD8[undef(addrspace=2)](nontemporal)(derenceable)(invariant)> t0, t2,
undef:i64
		  t2: i64,ch = CopyFromReg t0, Register:i64 %vreg2

		  t4: i64 = undef
	      t37: i64 = shl t35, Constant:i32<7>
		t35: i64 = add t29, t769
		  t29: i64 = add t929, t28


		  t769: i64 = zero_extend t768

		t101: i32 = Constant<7>
	    t61: i64 = Constant<64>
	  t4: i64 = undef
	t87: i32 = Constant<0>
      t881: i32 = extract_vector_elt t842, Constant:i32<1>
	t842: v4i32,ch = load<LD16[%arrayidx(addrspace=1)+64](align=64)> t112,
undef:i64
	  t62: i64 = add t38, Constant:i64<64>
	    t38: i64 = add t773, t37
	      t773: i64 = bitcast t772
		t772: v2i32,ch =
load<LD8[undef(addrspace=2)](nontemporal)(derenceable)(invariant)> t0, t2,
undef:i64
		  t2: i64,ch = CopyFromReg t0, Register:i64 %vreg2

		  t4: i64 = undef
	      t37: i64 = shl t35, Constant:i32<7>
		t35: i64 = add t29, t769
		  t29: i64 = add t929, t28


		  t769: i64 = zero_extend t768

		t101: i32 = Constant<7>
	    t61: i64 = Constant<64>
	  t4: i64 = undef
	t89: i32 = Constant<1>
  t771: f64 = bitcast t770
    t770: v2i32,ch =
load<LD8[undef(addrspace=2)](nontemporal)(dereferenceabinvariant)> t0, t8,
undef:i64
      t8: i64 = add t2, Constant:i64<8>
	t2: i64,ch = CopyFromReg t0, Register:i64 %vreg2
	  t1: i64 = Register %vreg2
	t7: i64 = Constant<8>
      t4: i64 = undef


This comes from a trivial OpenCL kernel:


__kernel void powGrid(__global KEDFOCLV *data, const double exponent){	       

    const size_t idx = get_global_id(0);
    data[idx] = pow(data[idx], exponent);
}

where KEDFOCLV is defined as double16.

Current benchmarks/clpeak hangs (reproduced by others on Polaris cards
additionally to my Carrizo). The latest github version of clpeak (see D10917)
crashes with an assertion in LLVM 4.0 on, at least, Carrizo and Polaris.

$ clpeak

Platform: Clover
  Device: AMD CARRIZO (DRM 3.8.0 / 12.0-CURRENT, LLVM 4.0.1)
    Driver version  : 17.1.0 (FreeBSD)
    Compute units   : 8
    Clock frequency : 800 MHz
Assertion failed: (DelayedTypos.empty() && "Uncorrected typos!"), function
~Sema, file
/wrkdirs/usr/ports/devel/llvm40/work/llvm-4.0.1rc1.src/tools/clang/lib/Sema/Sem
a.cpp,
line 317.
Abort trap (core dumped)