← Back to branch summary

~ubuntu-branches/ubuntu/trusty/libthrust/trusty

« back to all changes in this revision

Viewing changes to detail/backend/cuda/sort.inl

  • Committer: Package Import Robot
  • Author(s): Andreas Beckmann
  • Date: 2012-02-17 12:05:55 UTC
  • mfrom: (1.1.2)
  • Revision ID: package-import@ubuntu.com-20120217120555-0dg60rx22gy6nlg2
Tags: 1.5.1-1
* New upstream release.
* Set Maintainer to Debian NVIDIA Maintainers and move myself to Uploaders.

Show diffs side-by-side

added added

removed removed

Lines of Context:
detail/backend/cuda/sort.inl
30
30
#include <thrust/sequence.h>
31
31
#include <thrust/iterator/iterator_traits.h>
32
32
#include <thrust/detail/uninitialized_array.h>
 
33
#include <thrust/detail/trivial_sequence.h>
33
34
 
34
35
/*
35
36
 *  This file implements the following dispatch procedure for cuda::stable_sort()
269
270
                       StrictWeakOrdering comp,
270
271
                       thrust::detail::true_type)
271
272
    {
272
 
        // CUDA path for thrust::stable_sort with primitive keys
273
 
        // (e.g. int, float, short, etc.) and a less<T> or greater<T> comparison
274
 
        // method is implemented with stable_radix_sort
275
 
        thrust::detail::backend::cuda::detail::stable_radix_sort(first, last);
 
273
         // ensure sequence has trivial iterators
 
274
         thrust::detail::trivial_sequence<RandomAccessIterator> keys(first, last);
 
275
        
 
276
         // CUDA path for thrust::stable_sort with primitive keys
 
277
         // (e.g. int, float, short, etc.) and a less<T> or greater<T> comparison
 
278
         // method is implemented with stable_radix_sort
 
279
         thrust::detail::backend::cuda::detail::stable_radix_sort(keys.begin(), keys.end());
 
280
        
 
281
         // copy results back, if necessary
 
282
         if(!thrust::detail::is_trivial_iterator<RandomAccessIterator>::value)
 
283
             thrust::copy(keys.begin(), keys.end(), first);
276
284
       
277
 
        // if comp is greater<T> then reverse the keys
278
 
        typedef typename thrust::iterator_traits<RandomAccessIterator>::value_type KeyType;
279
 
        const static bool reverse = thrust::detail::is_same<StrictWeakOrdering, typename thrust::greater<KeyType> >::value;
 
285
         // if comp is greater<T> then reverse the keys
 
286
         typedef typename thrust::iterator_traits<RandomAccessIterator>::value_type KeyType;
 
287
         const static bool reverse = thrust::detail::is_same<StrictWeakOrdering, typename thrust::greater<KeyType> >::value;
280
288
 
281
 
        if (reverse)
282
 
          thrust::reverse(first, last);
 
289
         if (reverse)
 
290
           thrust::reverse(first, last);
283
291
    }
284
292
    
285
293
    template<typename RandomAccessIterator,
330
338
          thrust::reverse(values_first, values_first + (keys_last - keys_first));
331
339
        }
332
340
 
333
 
        thrust::detail::backend::cuda::detail::stable_radix_sort_by_key(keys_first, keys_last, values_first);
 
341
        // ensure sequences have trivial iterators
 
342
        thrust::detail::trivial_sequence<RandomAccessIterator1> keys(keys_first, keys_last);
 
343
        thrust::detail::trivial_sequence<RandomAccessIterator2> values(values_first, values_first + (keys_last - keys_first));
 
344
 
 
345
        thrust::detail::backend::cuda::detail::stable_radix_sort_by_key(keys.begin(), keys.end(), values.begin());
 
346
 
 
347
        // copy results back, if necessary
 
348
        if(!thrust::detail::is_trivial_iterator<RandomAccessIterator1>::value)
 
349
            thrust::copy(keys.begin(), keys.end(), keys_first);
 
350
        if(!thrust::detail::is_trivial_iterator<RandomAccessIterator2>::value)
 
351
            thrust::copy(values.begin(), values.end(), values_first);
334
352
        
335
353
        if (reverse)
336
354
        {