9prady9 · GitHub

@9prady9

WilliamTambellini

umar456

@9prady9

@9prady9

@9prady9 9prady9 changed the title Use static CUDA upstream libs and other changes to reduce global constant memory usage Use static CUDA upstream libs on Unix

Mar 14, 2020

@9prady9

@9prady9

umar456

umar456 previously approved these changes Mar 24, 2020

Instead of creating a static library out of all separate instantiations
of thrust_sort_by_key sources, we now directly embed sources
generated(using cmake's configure_file command) into afcuda target.
This also fixed separable compilation.
Prior to this change, separate compilation failed (related to cuda device
linking - undefined references). I tried to fix that problem, but
couldn't get a break through. However, I realized that just directly
using the generated sources with afcuda target will do the job without
any additional static library.

@9prady9

@9prady9

Open

@9prady9

thrust::stable_sort_by_key has known issue with device linking. The code
crashes with cudaInvalidValueError. It works as expected without any
changes with or without separable compilation otherwise.
https://github.com/thrust/thrust/wiki/Debugging#known-issues
https://github.com/thrust/thrust/blob/master/doc/changelog.md#known-issues-2
The above documents mention a known issue with device linking and thrust.
Although the documents say it happens in debug mode(with -G flag), I noticed
similar crashes in release configuration too in ArrayFire.
Due to the above issue, I have separated out the relevant source files
(fft,blas,sparse and solver) which require device linking into separate
static library. Once separated into a separate static library, sort_by_key
and all the other unit tests that use it are running as expected without
any crashes.

@9prady9

@9prady9

@9prady9

@9prady9

9prady9 deleted the use_static_culibs branch

April 1, 2020 04:08

Closed

Merged

Merged

2 tasks

Read the original on github.com ↗