CUDA를 사용하여 트리 구조체를 전달하는 방법

Question

간단한 C 트리를 만들고 GPU로 이식 할 계획입니다. 지금, 또는 CUDA에서 구조를 사용하는 방법, 나는 장치의 트리를 생성하는 방법 포트 CUDA의 코드를에 궁금

typedef struct node{ 

    short int final; // 2 byte 
    char number; // 1 byte 
    struct node *child[2]; // 8 * NUM_SIZE byte 

}node; 

:

내 트리의 구조는 다음과 같다? 또는 호스트에서 트리를 생성하여 장치에 전달할 수 있습니까?

디바이스에서 트리를 생성하는 것이 가장 좋은 방법 인 것 같습니다. 디바이스에서 구조를 사용하는 방법을 정확히 모르겠습니다.

감사

Answer 1

당신은 문제의이 종류에 사용할 수있는 세 가지 기본 방법이 있습니다

1. 호스트 API를 사용하여 트리에 대한 전역 메모리를 할당하고 호스트에 트리를 구성은, 그 복사 트리를 장치로 보냅니다.
2. 호스트 API를 사용하여 트리에 대한 전역 메모리를 할당하고
3. 장치의 트리에 대한 런타임 힙 메모리를 할당 장치에 트리를 구성하고 가장이다 장치

에 트리를 구성 적합한 사용 사례와 사용하는 GPU에 따라 다릅니다. (1) 호스트 메모리에 디바이스 트리 사본을 보유해야하고, (2) 디바이스 코드로 구현하기가 매우 어려운 동기화가 필요할 수 있으며, (3) 디바이스의 성능이 제한적이라는 단점이 있습니다 호스트 힙 API를 사용하여 직접 액세스 할 수없는 트리를 생성합니다.

마지막으로 포인터를 기반으로하는 트리는 일반적으로 GPU에 적합하지 않거나 GPU에 적합하지 않습니다 질문에 설명 된 것과 같은 구조를 사용하기 전에 대체 데이터 구조와 알고리즘을 고려해야합니다.

---

편집 : 그래서 여기 당신이 그것을 수행하는 방법의 몇 가지 아이디어를 얻을 볼 수있는 절대적으로 사소한 예입니다

분명히 이러한 모든 옵션은 당신을 위해 너무 어려운 증명하고

. 먼저 코드 : 여기

#include <iostream> 

struct __align__(8) node{ 
    char val; 
    struct node *child; 
}; 

__global__ 
void kernel(node * tree, char *out, int n) 
{ 
    node *p = tree; 
    int i=0; 
    while(p->val != 0) { 
    out[i++] = p->val; 
    p = p->child; 
    } 
} 

int main(void) 
{ 
    const int n = 15; 
    char data[n] = "tietamattomana"; 
    node tree[n]; 

    node * tree_d; 
    char * output_d; 
    cudaMalloc((void **)&tree_d, n * sizeof(node)); 
    cudaMalloc((void **)&output_d, n * sizeof(char)); 

    node * p = tree_d; 
    for(int i=0; i<n; i++) { 
     tree[i].val = data[i]; 
     tree[i].child = (++p); 
    } 

    cudaMemcpy(tree_d, tree, n * sizeof(node), cudaMemcpyHostToDevice); 
    kernel<<<1,1>>>(tree_d, output_d, n); 

    char output[n]; 
    cudaMemcpy(output, output_d, n * sizeof(char), cudaMemcpyDeviceToHost); 
    for(int i=0; i<n; i++) { 
     std::cout << output[i]; 
    } 
    std::cout << std::endl; 

    return 0; 
} 

, 방금 호스트에서 간단한 연결리스트를 채워 각 목록의 노드가 호스트 문자열에서 값을 유지하는 장치로 복사 한. 단일 GPU 스레드는 목록을 처음부터 끝까지 추적하여 각 노드의 값을 읽고 출력 배열에 저장합니다. , 즉 출력 배열의 내용을 모든 작업, 출력 배열 다시 GPU에서 호스트 사본을 확인하고 에코하려면

$ nvcc -arch=sm_30 -Xptxas="-v" tree.cu 
ptxas info : 0 bytes gmem 
ptxas info : Compiling entry function '_Z6kernelP4nodePci' for 'sm_30' 
ptxas info : Function properties for _Z6kernelP4nodePci 
    0 bytes stack frame, 0 bytes spill stores, 0 bytes spill loads 
ptxas info : Used 10 registers, 340 bytes cmem[0] 

$ ./a.out 
tietamattomana 

는 아마도 이것은 적어도 당신은 당신 것입니다 무엇에 길을 시작합니다 예를 들어, 내가 언급 한 다른 가능성들이 실제로 작용할 수도 있고 내가 언급 한 단점들이 어떻게 작용할지를 밝힐 수도있다.