CUDA toàn cầu (như trong C) mảng động được phân bổ cho bộ nhớ thiết bị

Question

Vì vậy, tôi đang cố gắng viết một số mã sử dụng kiến ​​trúc CUDA của Nvidia. Tôi nhận thấy rằng việc sao chép đến và từ thiết bị thực sự làm tổn thương hiệu suất tổng thể của tôi, vì vậy bây giờ tôi đang cố di chuyển một lượng lớn dữ liệu vào thiết bị.

Vì dữ liệu này được sử dụng trong nhiều chức năng, tôi muốn nó là toàn cầu. Có, tôi có thể truyền con trỏ xung quanh, nhưng tôi thực sự muốn biết làm thế nào để làm việc với globals trong trường hợp này.

Vì vậy, tôi có các chức năng thiết bị muốn truy cập mảng được phân bổ thiết bị.

Lý tưởng nhất, tôi có thể làm một cái gì đó như:

__device__ float* global_data; 

main() 
{ 
    cudaMalloc(global_data); 
    kernel1<<<blah>>>(blah); //access global data 
    kernel2<<<blah>>>(blah); //access global data again 
} 

Tuy nhiên, tôi havent đã tìm ra cách để tạo ra một mảng động. Tôi đã tìm ra một công việc xung quanh bằng cách tuyên bố mảng như sau:

__device__ float global_data[REALLY_LARGE_NUMBER]; 

Và trong khi đó không yêu cầu cuộc gọi cudaMalloc, tôi muốn tiếp cận phân bổ động.

Answer 1

Điều gì đó giống như vậy có thể hoạt động.

#include <algorithm> 

#define NDEBUG 
#define CUT_CHECK_ERROR(errorMessage) do {         \ 
     cudaThreadSynchronize();           \ 
     cudaError_t err = cudaGetLastError();        \ 
     if(cudaSuccess != err) {           \ 
        fprintf(stderr, "Cuda error: %s in file '%s' in line %i : %s.\n", \ 
              errorMessage, __FILE__, __LINE__, cudaGetErrorString(err));\ 
        exit(EXIT_FAILURE);             \ 
       } } while (0) 


__device__ float *devPtr; 

__global__ 
void kernel1(float *some_neat_data) 
{ 
    devPtr = some_neat_data; 
} 

__global__ 
void kernel2(void) 
{ 
    devPtr[threadIdx.x] *= .3f; 
} 


int main(int argc, char *argv[]) 
{ 
    float* otherDevPtr; 
    cudaMalloc((void**)&otherDevPtr, 256 * sizeof(*otherDevPtr)); 
    cudaMemset(otherDevPtr, 0, 256 * sizeof(*otherDevPtr)); 

    kernel1<<<1,128>>>(otherDevPtr); 
    CUT_CHECK_ERROR("kernel1"); 

    kernel2<<<1,128>>>(); 

    CUT_CHECK_ERROR("kernel2"); 

    return 0; 
} 

Cho nó một vòng xoáy.

Answer 2

Dành một chút thời gian tập trung vào tài liệu phong phú do NVIDIA cung cấp.

Từ Programming Guide:

float* devPtr; 
cudaMalloc((void**)&devPtr, 256 * sizeof(*devPtr)); 
cudaMemset(devPtr, 0, 256 * sizeof(*devPtr)); 

Đó là một ví dụ đơn giản về cách cấp phát bộ nhớ. Bây giờ, trong hạt nhân của bạn, bạn nên chấp nhận một con trỏ đến một phao như vậy:

__global__ 
void kernel1(float *some_neat_data) 
{ 
    some_neat_data[threadIdx.x]++; 
} 

__global__ 
void kernel2(float *potentially_that_same_neat_data) 
{ 
    potentially_that_same_neat_data[threadIdx.x] *= 0.3f; 
} 

Vì vậy, bây giờ bạn có thể gọi họ như vậy:

float* devPtr; 
cudaMalloc((void**)&devPtr, 256 * sizeof(*devPtr)); 
cudaMemset(devPtr, 0, 256 * sizeof(*devPtr)); 

kernel1<<<1,128>>>(devPtr); 
kernel2<<<1,128>>>(devPtr); 

Theo dữ liệu này được sử dụng trong nhiều chức năng, tôi muốn nó là toàn cầu.

Có một vài lý do chính đáng để sử dụng hình cầu. Điều này chắc chắn không phải là một. Tôi sẽ để nó như một bài tập để mở rộng ví dụ này để bao gồm việc chuyển "devPtr" sang phạm vi toàn cầu.

EDIT:

Ok, vấn đề cơ bản là thế này: hạt nhân của mình chỉ có thể truy cập bộ nhớ điện thoại và các con trỏ toàn cầu phạm vi duy nhất mà họ có thể sử dụng là những GPU. Khi gọi một hạt nhân từ CPU của bạn, đằng sau hậu trường những gì xảy ra là các con trỏ và nguyên thủy được sao chép vào thanh ghi GPU và/hoặc bộ nhớ chia sẻ trước khi hạt nhân được thực thi.

Vì vậy, gần nhất tôi có thể đề xuất là: sử dụng cudaMemcpyToSymbol() để đạt được mục tiêu của bạn. Nhưng, ở chế độ nền, hãy xem xét một cách tiếp cận khác có thể là Điều Đúng.

#include <algorithm> 

__constant__ float devPtr[1024]; 

__global__ 
void kernel1(float *some_neat_data) 
{ 
    some_neat_data[threadIdx.x] = devPtr[0] * devPtr[1]; 
} 

__global__ 
void kernel2(float *potentially_that_same_neat_data) 
{ 
    potentially_that_same_neat_data[threadIdx.x] *= devPtr[2]; 
} 


int main(int argc, char *argv[]) 
{ 
    float some_data[256]; 
    for (int i = 0; i < sizeof(some_data)/sizeof(some_data[0]); i++) 
    { 
     some_data[i] = i * 2; 
    } 
    cudaMemcpyToSymbol(devPtr, some_data, std::min(sizeof(some_data), sizeof(devPtr))); 
    float* otherDevPtr; 
    cudaMalloc((void**)&otherDevPtr, 256 * sizeof(*otherDevPtr)); 
    cudaMemset(otherDevPtr, 0, 256 * sizeof(*otherDevPtr)); 

    kernel1<<<1,128>>>(otherDevPtr); 
    kernel2<<<1,128>>>(otherDevPtr); 

    return 0; 
} 

Đừng quên '--host-compilation = C++' cho ví dụ này.

Answer 3

Erm, chính xác là vấn đề di chuyển devPtr đến phạm vi toàn cầu là vấn đề của tôi.

Tôi có triển khai thực hiện chính xác điều đó, với hai hạt nhân có con trỏ tới dữ liệu được truyền vào.Tôi rõ ràng không muốn vượt qua những con trỏ đó.

Tôi đã đọc tài liệu khá chặt chẽ, và truy cập diễn đàn nvidia (và google tìm kiếm một giờ hoặc lâu hơn), nhưng tôi chưa tìm thấy triển khai mảng thiết bị động toàn cầu thực sự chạy (tôi đã thử một số biên dịch và sau đó thất bại theo những cách mới và thú vị).

Answer 4

kiểm tra các mẫu đi kèm với SDK. Nhiều trong số những dự án mẫu đó là một cách tốt để học bằng ví dụ.

Answer 5

Tôi đã tiếp tục và thử giải pháp cấp phát một con trỏ tạm thời và chuyển nó tới một hàm toàn cục đơn giản tương tự như kernel1.

Các tin tốt là nó không làm việc :)

Tuy nhiên, tôi nghĩ rằng nó lẫn lộn trình biên dịch như tôi bây giờ có được "tư vấn: Không thể nói điểm gì con trỏ đến, giả sử không gian bộ nhớ toàn cầu" bất cứ khi nào tôi cố gắng truy cập dữ liệu toàn cầu. May mắn thay, giả định xảy ra là đúng, nhưng các cảnh báo rất khó chịu.

Dù sao, để ghi lại - tôi đã xem xét nhiều ví dụ và đã chạy qua các bài tập nvidia nơi mà điểm là để có được đầu ra để nói "Đúng!". Tuy nhiên, tôi chưa xem xét tất cả trong số đó. Nếu có ai biết về một ví dụ về sdk nơi họ làm phân bổ bộ nhớ thiết bị toàn cầu động, tôi vẫn muốn biết.

Answer 6

Vì dữ liệu này được sử dụng trong nhiều chức năng, tôi muốn nó là toàn cục.

-

Có vài lý do tốt để sử dụng globals. Điều này chắc chắn không phải là một. Tôi sẽ để nó như một bài tập để mở rộng ví dụ này để bao gồm việc di chuyển "devPtr" sang phạm vi toàn cầu.

Điều gì xảy ra nếu hạt nhân hoạt động trên cấu trúc const lớn bao gồm mảng? Sử dụng cái gọi là bộ nhớ liên tục không phải là một lựa chọn, bởi vì nó rất hạn chế về kích thước .. vì vậy sau đó bạn phải đặt nó trong bộ nhớ toàn cầu ..?