| Bug ID | 99552 |
|---|---|
| Summary | Make Advanced Simulation Library (ASL) OpenCL GPU support wor= k on Clover and RadeonSI |
| Product | Mesa |
| Version | git |
| Hardware | All |
| OS | All |
| Status | NEW |
| Severity | normal |
| Priority | medium |
| Component | Drivers/Gallium/radeonsi |
| Assignee | dri-devel@lists.freedesktop.org |
| Reporter | vedran@miletic.net |
| QA Contact | dri-devel@lists.freedesktop.org |
Advanced Simulation Library (ASL) is a partial differential eq=
uation solver
that supports running with OpenCL. When using it on Mesa/radeonsi/LLVM/AMDG=
PU,
test/testACL/testKernel.cc [1] partially works but ultimately fails with:
$ ./testKernel
Test of "copy" function... Ok
Test of Kernel with double... Ok
Test of KernelSIMD... Ok
Test of KernelSIMDUA... Ok
Test of kernel with PrivateVariable... Ok
Test of kernel with PrivateArray...
BUILD LOG
************************************************
<unknown>:0:0: in function compute_4 void (i32 addrspace(1)*, float
addrspace(1)*): invalid addrspacecast
************************************************
KERNEL SOURCE CODE
------------------------------------------------
#pragma OPENCL EXTENSION cl_khr_fp64 : disable
__kernel void compute_4(__global int *a_i1,
__global float *a_f8)
{
uint index =3D get_global_id(0);
float pa_f1[6] =3D {-9, 2, 0, 15, 1, 3};
(a_f8[index]=3Dpa_f1[a_i1[index]]);
}
------------------------------------------------
terminate called after throwing an instance of 'std::logic_error'
what(): ASL ERROR: Program::build() (-11).
Aborted (core dumped)
[1]
https://github.com/AvtechScientific/ASL/blob/master/test/te=
stACL/testKernel.cc
| What | Removed | Added |
|---|---|---|
| Blocks | 99553 |
How to test:
$ mkdir build; cd build
$ cmake -DWITH_TESTS=3DON ..
$ make -j
$ make test
Right now, Mesa/LLVM latest git pulled half an hour ago I get:
Running tests...
Test project /users/miletivn/workspace/ASL/build
Start 1: testABDFormat
1/10 Test #1: testABDFormat .................... Passed 0.07 sec
Start 2: testVectorOfElements
2/10 Test #2: testVectorOfElements ............. Passed 5.27 sec
Start 3: testMatrixOfElements
3/10 Test #3: testMatrixOfElements ............. Passed 7.89 sec
Start 4: testKernel
4/10 Test #4: testKernel .......................***Exception: Other 13.94=
sec
Start 5: testOperators
5/10 Test #5: testOperators .................... Passed 7.88 sec
Start 6: testKernelMerger
6/10 Test #6: testKernelMerger ................. Passed 2.65 sec
Start 7: testPrivateVar
7/10 Test #7: testPrivateVar ...................***Exception: Other 3.22=
sec
Start 8: testASLData
8/10 Test #8: testASLData ...................... Passed 13.72 sec
Start 9: testDistanceFunction
9/10 Test #9: testDistanceFunction ............. Passed 42.43 sec
Start 10: testReductionFunction
10/10 Test #10: testReductionFunction ............ Passed 29.24 sec
80% tests passed, 2 tests failed out of 10
Label Time Summary:
DoublePrecision =3D 85.62 sec (3 tests)
Performance =3D 3.22 sec (1 test)
Total Test time (real) =3D 126.34 sec
The following tests FAILED:
4 - testKernel (OTHER_FAULT)
7 - testPrivateVar (OTHER_FAULT)
Errors while running CTest
Makefile:61: recipe for target 'test' failed
make: *** [test] Error 8
(In reply to Vedran Mileti=C4=87 from comment #1) > Start 7: testPrivateVar > 7/10 Test #7: testPrivateVar ...................***Exception: Other = 3.22 > sec Adding native_rsqrt fixes this, patch will appear here: http:= //lists.llvm.org/pipermail/libclc-dev/2017-February/
| What | Removed | Added |
|---|---|---|
| Resolution | --- | FIXED |
| Status | NEW | RESOLVED |
Fixed https://revi= ews.llvm.org/D27283 https= ://reviews.llvm.org/rL294786