Does CUDA support recursion?

RecursionCuda

Recursion Problem Overview


Does CUDA support recursion?

Recursion Solutions


Solution 1 - Recursion

It does on NVIDIA hardware supporting compute capability 2.0 and CUDA 3.1:

> New language features added to CUDA C > / C++ include: > > Support for function > pointers and recursion make it easier > to port many existing algorithms to > Fermi GPUs

http://developer.nvidia.com/object/cuda_3_1_downloads.html

Function pointers: http://developer.download.nvidia.com/compute/cuda/sdk/website/CUDA_Advanced_Topics.html#FunctionPointers

Recursion: I can't find a code sample on NVIDIA's website, but on the forum someone post this:

__device__ int fact(int f)
{
  if (f == 0)
    return 1;
  else
    return f * fact(f - 1);
}

Solution 2 - Recursion

Yes, see the NVIDIA CUDA Programming Guide:

> device functions only support recursion in device code compiled for devices of compute capability 2.0.

You need a Fermi card to use them.

Solution 3 - Recursion

Even though it only supports recursion for specific chips, you can sometimes get away with "emulated" recursion: see how I used compile-time recursion for my CUDA raytracer.

Solution 4 - Recursion

In CUDA 4.1 release CUDA supports recursion only for _device_ function but not for _global_ function.

Solution 5 - Recursion

Only after 2.0 compute capability on compatible devices

Solution 6 - Recursion

Sure it does, but it requires the Kepler architecture to do so. Check out their latest example on the classic quick sort.

http://blogs.nvidia.com/2012/09/how-tesla-k20-speeds-up-quicksort-a-familiar-comp-sci-code/

As far as i know, only latest Kepler GK110 supports dynamic parallelism, which allow this kind of recursive call and spawning of new threads within the kernel. Before Kepler GK110, it was not possible. And note that not all Kepler architecture supports this, only GK110 does.

If you need recursion, you probably need the Tesla K20. I'm not sure if Fermi does supports it,never read of it. :
But Kepler sure does. =)

Solution 7 - Recursion

Any recursive algorithm can be implemented with a stack and a loop. It's way more of a pain, but if you really need recursion, this can work.

Solution 8 - Recursion

CUDA 3.1 supports recursion

Solution 9 - Recursion

If your algorithm invovles alot of recursions, then support or not, it is not designed for GPUs, either redesign your algorthims or get a better CPU, either way it will be better (I bet in many cases, maginitudes better) then do recurisons on GPUs.

Solution 10 - Recursion

Yeah, it is supported on the actual version. But despite the fact it is possible to execute recursive functions, you must have in mind that the memory allocation from the execution stack cannot be predicted (the recursive function must be executed in order to know the true depth of the recursion), so your stack could result being not enough for your purposes and it could need a manual increment of the default stack size

Solution 11 - Recursion

Yes, it does support recursion. However, it is not a good idea to do recursion on GPU. Because each thread is going to do it.

Solution 12 - Recursion

Tried just now on my pc with a NVIDIA GPU with 1.1 Compute capability. It says recursion not yet supported. So its not got anything to do with the runtime but the hardware itself

Attributions

All content for this solution is sourced from the original question on Stackoverflow.

The content on this page is licensed under the Attribution-ShareAlike 4.0 International (CC BY-SA 4.0) license.

Content TypeOriginal AuthorOriginal Content on Stackoverflow
QuestionJuanPabloView Question on Stackoverflow
Solution 1 - RecursionStringerView Answer on Stackoverflow
Solution 2 - RecursionDr. SnoopyView Answer on Stackoverflow
Solution 3 - RecursionttsiodrasView Answer on Stackoverflow
Solution 4 - Recursionusername_4567View Answer on Stackoverflow
Solution 5 - RecursionArturo GarciaView Answer on Stackoverflow
Solution 6 - RecursionHong ZhouView Answer on Stackoverflow
Solution 7 - RecursiondicroceView Answer on Stackoverflow
Solution 8 - RecursionJan CView Answer on Stackoverflow
Solution 9 - Recursionuser0002128View Answer on Stackoverflow
Solution 10 - RecursionMr.PotatusVIIView Answer on Stackoverflow
Solution 11 - RecursionpalebluedotView Answer on Stackoverflow
Solution 12 - Recursionsp497View Answer on Stackoverflow