     }
   }
 };
-VariableUnitTest<TestRadixSortByKeyVariableBits, IntegralTypes> TestRadixSortByKeyVariableBitsInstance;
+VariableUnitTest<TestRadixSortByKeyVariableBits, unittest::type_list<unsigned int> > TestRadixSortByKeyVariableBitsInstance;
 
 
 //void TestRadixSortMemoryLeak(void)

  {
    thrust::host_vector<T>   h_keys = unittest::random_integers<T>(n);
    thrust::device_vector<T> d_keys = h_keys;
    
    thrust::host_vector<unsigned int>   h_values(n);
    thrust::device_vector<unsigned int> d_values(n);
    thrust::sequence(h_values.begin(), h_values.end());


//still need to do long/ulong and maybe double

typedef unittest::type_list<
#if !(defined(__GNUC__) && (__GNUC__ <= 4) && (__GNUC_MINOR__ <= 1))
// XXX GCC 4.1 miscompiles the char sorts with -O2 for some reason
                            char,
                            signed char,
                            unsigned char,
#endif
                            short,
                            unsigned short,
                            int,
                            unsigned int,
                            long,
                            unsigned long,
                            long long,
                            unsigned long long,
                            double> RadixSortKeyTypes;

template <typename T>
struct TestRadixSort
