Error when building Sofa's pluging "no instance of overloaded function "atomicAdd" matches the argument list" #2660
-
|
Hi, I have build sofa sucessfully. But for the some plugins, I encountered the following error. From the commit history, i guess it is related to the commit 1ae7fc6 but not sure. Thanks |
Beta Was this translation helpful? Give feedback.
Replies: 5 comments 5 replies
-
Note: I am not a CUDA expert at all |
Beta Was this translation helpful? Give feedback.
-
|
Hi, I have tried to turn the SOFACUDA_DOUBLE=OFF, it compiled successfully the previous error but resulted in the new error: |
Beta Was this translation helpful? Give feedback.
-
|
This seems to be related to your own SofaCUDASolvers repository @longhathuc to which we do not have access / history |
Beta Was this translation helpful? Give feedback.
-
|
Without looking at the code of this plugin SofaCUDASolvers we can see that it needs templated stuff on double (e.g |
Beta Was this translation helpful? Give feedback.
-
|
Very strange I tried the same builds in a laptop and it works fine. This might be a problem of my computer. I will check later. |
Beta Was this translation helpful? Give feedback.
-
|
Alright @longhathuc keep us updated please |
Beta Was this translation helpful? Give feedback.
-
|
I've managed to find the reason of this. Thus, the only workaround I've found (because the computer I use in the lab also have a GTX 9XX) is to either put SOFA_GPU_CUDA_DOUBLE to false, or just comment the part related to double in this particular component. In fact, this component was only compiled in float until @epernod 's commit on the 14th of January (1ae7fc6), until then, we were able to compile with cuda double enabled (the other components weren't using this atomicAdd function). So this may be prone to a discussion in the future (add a flag for this component in particular regarding the double instantiation) |
Beta Was this translation helpful? Give feedback.
-
|
@bakpaul I think you right. Here is a solution suggested by Nvidia:
#if __CUDA_ARCH__ < 600
__device__ double atomicAdd(double* address, double val)
{
unsigned long long int* address_as_ull =
(unsigned long long int*)address;
unsigned long long int old = *address_as_ull, assumed;
do {
assumed = old;
old = atomicCAS(address_as_ull, assumed,
__double_as_longlong(val +
__longlong_as_double(assumed)));
// Note: uses integer comparison to avoid hang in case of NaN (since NaN != NaN)
} while (assumed != old);
return __longlong_as_double(old);
}
#endifSource: https://docs.nvidia.com/cuda/cuda-c-programming-guide/index.html. Could you please try adding this code snippet if it fix your compilation? |
Beta Was this translation helpful? Give feedback.
-
|
Well, CUDA_ARCH is not defined and neither is __longlong_as_double . So it doesn't work as simply. And I would add that even if it compiles, if the GPU is not meant to use atomic methods in double, it will not work in the run-time. So this may not be a solution that fits our needs. Thankfully (we are not using this component in any way on our side, so it is not a big issue If it doesn't work... I will try to find a time next week to fix the compilation using this code, (I am on the end of my PhD so...), but again maybe a flag could be better for the people who want to have Double support on the other components while keeping this one as a float but only have a 9XX GPU... |
Beta Was this translation helpful? Give feedback.
-
|
Let us know if you managed to compile it while working similarly to |
Beta Was this translation helpful? Give feedback.
@bakpaul I think you right.
Here is a solution suggested by Nvidia: