Thrust launch failure

[FlorianReiss]

When using CUDA 11 and the associated Thrust version, issues start appearing in Amp3Body when
thrust::transform or thrust::transform_reduce are called.

RuntimeError: transform: failed to synchronize: cudaErrorLaunchFailure: unspecified launch failure
terminate called after throwing an instance of 'thrust::system::system_error'
  what():  CUDA free failed: cudaErrorLaunchFailure: unspecified launch failure

They can be fixe by adding thrust::seq (see also),
but this is not ideal and instead another error starts appearing.

terminate called after throwing an instance of 'thrust::system::system_error'
  what():  CUDA free failed: cudaErrorCudartUnloading: driver shutting down

This seems to be independent of the resonances included in Amp3Body, that is not only a simple example with the kMatrix fails, but also other resonances like RBW

to reproduce on sneezy:

ml load gcc/7.5/cuda/11.2
source /usr/local/anaconda3/etc/profile.d/conda.sh
build GooFit from branch freiss_daniel_CUDA11 with  cmake -S . -B build-cuda -DGOOFIT_FORCE_LOCAL_THRUST=OFF -DGOOFIT_CERNROOT=OFF -DGOOFIT_PYTHON=ON -DGOOFIT_KMATRIX=ON -DGOOFIT_EXAMPLES=OFF -DGOOFIT_TESTS=OFF   -DCMAKE_CUDA_FLAGS_DEBUG="-g -G" -DCMAKE_BUILD_TYPE=Debug
set PYTHONPATH

[FlorianReiss]

something which might help understand this issue: I tried using a simple function with thrust::transform, but only could get it to work when declaring a __host__ __device__ function, instead of a __device__ function

[FlorianReiss]

part of the solution might be to add thrust::device to the transform call

[FlorianReiss]

in the end, it seems like culprit was trying to compile with both -DCMAKE_BUILD_TYPE=Debug and -DCMAKE_CUDA_FLAGS_DEBUG="-g -G" flags. Just having one of them seems fine, while having both results in the issues described above