CUDA推力裝置strcmp

Question

我已經使下列代碼能夠用於使用THRUST排序方法對char *數組進行排序。出於某種原因，代碼每當它試圖比較字符串中的字符時都會中斷。

thrust::sort_by_key(deviceArrayToSort.begin(),deviceArrayToSort.end(),deviceArrayToSortRow.begin(),CharArrayCmp()); 

比較器是如下：

struct CharArrayCmp{ 
__host__ __device__ 
    bool operator()(const CharArr & o1, const CharArr & o2) { 
     return this->compare(o1.value,o1.length,o2.value,o2.length); 
    } 

    __host__ __device__ bool compare (const char * src, int lenSrc, const char * dst, int lenDest) 
{ 
    int end; 
    if(lenSrc > lenDest){ 
     end = lenDest; 
    }else{ 
     end = lenSrc; 
    } 
    for(int i = 0; i < end; i++){ 
     if(src[i] > dst[i]){ 
      return false; 
     } 
    } 
    if(lenSrc >= lenDest){ 
     return false; 
    } 
    return true; 
} 

};

當它試圖在這條線運行在設備上有一個錯誤：

if(src[i] != dst[i]) 

    thrust::system::cuda::detail::detail::launch_closure_by_value<thrust::system::cuda::detail:: 
detail::stable_sort_by_count_detail::stable_sort_by_count_closure<256u, 
thrust::detail::normal_iterator<thrust::pointer<unsigned int, 
thrust::system::cuda::detail::tag, thrust::use_default, thrust::use_default> >, 
thrust::detail::normal_iterator<thrust::pointer<unsigned int, 
thrust::system::cuda::detail::tag, thrust::use_default, thrust::use_default> >, 
    thrust::system::cuda::detail::temporary_indirect_ordering<thrust::system::cuda::detail::tag, 
thrust::detail::normal_iterator<thrust::device_ptr<CharArr> >, CharArrayCmp>::compare, 
thrust::system::cuda::detail::detail::statically_blocked_thread_array<256u> > > [0] [device 
0 (GK104)] (Signal) - CUDA_EXCEPTION_10:Device Illegal Address 

CUDA Thread (193,0,0) Block (22,0,0)  


All Kernel Threads (144 Blocks of 256 Threads) 

我很新的CUDA，所以我不確定我在做什麼錯，但感覺這應該是非常簡單。

這是CharArr結構：

typedef struct{ 

char * value; 
int length; 
} CharArr; 

最後這裏就是利用這個sort_by_key的代碼。我已確保傳遞給此信息是正確的。即arrayToSort和arrayToSortRow都是數組，分別是char *和long long，size是這兩個數組的大小。

void sortCharArrayStable(char ** arrayToSort, long long * arrayToSortRow,long long size){ 
std::cout <<"about to start LongIndex" <<std::endl; 

     thrust::host_vector<CharArr> hostToSort(size); 
     thrust::host_vector<long long> hostToSortRow(size); 
     for(int i =0; i < size; i++){ 
      CharArr sortRow; 
      if(arrayToSort[i] == 0x0){ 
       sortRow.length = 0; 
       sortRow.value = ""; 
       std::cout<<"Had an error on row "<< arrayToSortRow[i]<<" when making column array for sortCharArrayStable"<<std::endl; 

      }else{ 

       sortRow.length = strlen(arrayToSort[i]); 
       sortRow.value = arrayToSort[i]; 
      } 



      hostToSort[i] = sortRow; 
      hostToSortRow[i] = arrayToSortRow[i]; 
     } 
     thrust::device_vector<CharArr> deviceArrayToSort = hostToSort;// (arrayToSort,arrayToSort + size); 
     thrust::device_vector<long long> deviceArrayToSortRow = hostToSortRow; 




thrust::stable_sort_by_key(deviceArrayToSort.begin(),deviceArrayToSort.end(),deviceArrayToSortRow.begin(),CharArrayCmp()); 

     //copy the contents back into our original array to sort now sorted 
     hostToSort = deviceArrayToSort; 
     for(int i =0; i < size; i++){ 
      arrayToSort[i] = hostToSort[i].value; 
     } 
     arrayToSortRow); 


thrust::copy(deviceArrayToSortRow.begin(),deviceArrayToSortRow.end(),arrayToSortRow); 


} 

這裏是一個編譯例子ocurring問題的一個完整的例子：

#include <thrust/device_vector.h> 
#include <thrust/host_vector.h> 
#include <thrust/reduce.h> 
#include <thrust/device_vector.h> 
#include <thrust/host_vector.h> 
#include <thrust/sort.h> 

#include <thrust/reduce.h> 


typedef struct{ 

    char * value; 
    int length; 
} CharArr; 


struct CharArrayCmp{ 
    __host__ __device__ 
     bool operator()(const CharArr & o1, const CharArr & o2) { 
      return this->compare(o1.value,o1.length,o2.value,o2.length); 
     } 

    __host__ __device__ bool compare (const char * src, int lenSrc, const char * dst, int lenDest) 
    { 
     int end; 
     if(lenSrc > lenDest){ 
      end = lenDest; 
     }else{ 
      end = lenSrc; 
     } 
     for(int i = 0; i < end; i++){ 
      if(src[i] > dst[i]){ 
       return false; 
      } 
     } 
     if(lenSrc >= lenDest){ 
      return false; 
     } 
     return true; 
    } 
}; 


void sortCharArray(char ** arrayToSort, long long * arrayToSortRow,long long size){ 
    std::cout <<"about to start LongIndex" <<std::endl; 

      thrust::host_vector<CharArr> hostToSort(size); 
      thrust::host_vector<long long> hostToSortRow(size); 
      for(int i =0; i < size; i++){ 
       CharArr sortRow; 
       sortRow.value = arrayToSort[i]; 
       sortRow.length = strlen(arrayToSort[i]); 
       hostToSort[i] = sortRow; 
       hostToSortRow[i] = arrayToSortRow[i]; 
      } 
      thrust::device_vector<CharArr> deviceArrayToSort = hostToSort;// (arrayToSort,arrayToSort + size); 
      thrust::device_vector<long long> deviceArrayToSortRow = hostToSortRow;//(arrayToSortRow,arrayToSortRow + size); 

      // thrust::sort(deviceArrayToSort.begin(),deviceArrayToSort.end()); 
      thrust::sort_by_key(deviceArrayToSort.begin(),deviceArrayToSort.end(),deviceArrayToSortRow.begin(),CharArrayCmp()); 

      //copy the contents back into our original array to sort now sorted 
      hostToSort = deviceArrayToSort; 
      for(int i =0; i < size; i++){ 
       arrayToSort[i] = hostToSort[i].value; 
      } 
      thrust::copy(deviceArrayToSortRow.begin(),deviceArrayToSortRow.end(),arrayToSortRow); 


} 

int main() 
{ 
    char ** charArr = new char*[10]; 

    charArr[0] = "zyxw"; 
    charArr[1] = "abcd"; 
    charArr[2] = "defg"; 
    charArr[3] = "werd"; 
    charArr[4] = "aasd"; 
    charArr[5] = "zwedew"; 
    charArr[6] = "asde"; 
    charArr[7] = "rurt"; 
    charArr[8] = "ntddwe"; 
    charArr[9] = "erbfde"; 

    long long * rows = new long long[10]; 
    for(int i = 0; i < 10;i++){ 
     rows[i] = i; 
    } 

    sortCharArray(charArr,rows,10); 

    for(int i = 0; i < 10; i++){ 
     std::cout<<"Row is "<<rows[i]<<" String is "<<charArr[i]<<std::endl; 

    } 
} 

Answer 1

這是行不通的：

thrust::device_vector<CharArr> deviceArrayToSort = hostToSort;// (arrayToSort,arrayToSort + size); 
    thrust::device_vector<long long> deviceArrayToSortRow = hostToSortRow; 

    thrust::stable_sort_by_key(deviceArrayToSort.begin(),deviceArrayToSort.end(),deviceArrayToSortRow.begin(),CharArrayCmp()); 

每個CharArr對象hostToSort包含一個指針點到主機內存位置。當您將該指針複製到設備矢量時，該指針的數值爲，不變爲。如果您嘗試在設備代碼中取消引用此（僞造）指針，它將會像您所看到的那樣失敗。

您將需要經過deviceArrayToSort設備向量，併爲它的每個CharArr對象，則需要調整其value指針指向設備存儲器的有效位置，這大概也是每個起始地址字符串進行排序。