Волгоградский филиал

ZEnable – имеет значение true, если используется Z-бу­фер, либо false, если он не используется. При физическом моделировании Z-буфер не используется, так что ZEnable = false;

CullMode – устанавливает режим отсечения невидимых треугольников. При физическом моделировании отсечения треугольников нет, так что CullMode = none.

 

Приложение 2
Процедуры CUDA, исполняемые на CPU

// Файл matrixMul.cu [68]

 

// Copyright 1993-2007 NVIDIA Corporation. All rights reserved.

 

/* Умножение матриц: C = A * B. Вызывающий код на CPU.

*

* В этом примере реализовано умножение матриц как приведено в главе 7

* документации CUDA Programming Guide [68]. Эта реализация предназначена

* для изучения принципов программирования на CUDA, поэтому эффективность

* принесена в жертву ясности изложения.

*

* Высокопроизводительная версия умножения матриц реализована в библиотеке CUBLAS

*/

 

#include < stdlib.h>

#include < stdio.h>

#include < string.h>

#include < math.h>

#include < cutil.h>

 

#include < matrixMul_kernel.cu>

 

////////////////////////////////////////////////////////////////////////////////

// предварительное объявление функций

void runTest(int argc, char** argv);

void randomInit(float*, int);

void printDiff(float*, float*, int, int);

 

/* Процедура, которая для сравнения перемножает матрицы на центральном процессоре: */

 

extern " C" void computeGold(float*, const float*, const float*, unsigned int, unsigned int, unsigned int);

 

int main(int argc, char** argv)

{

runTest(argc, argv);

CUT_EXIT(argc, argv);

}

 

void runTest(int argc, char** argv)

{

CUT_DEVICE_INIT();

 

// инициализация генератора псевдослучайных чисел rand()

srand(2006);

 

// выделение памяти для матриц A и B

unsigned int size_A = WA * HA;

unsigned int mem_size_A = sizeof(float) * size_A;

float* h_A = (float*) malloc(mem_size_A);

unsigned int size_B = WB * HB;

unsigned int mem_size_B = sizeof(float) * size_B;

float* h_B = (float*) malloc(mem_size_B);

 

// заполнение исходных матриц псевдослучайными числами

randomInit(h_A, size_A);

randomInit(h_B, size_B);

 

// выделение памяти на GPU

float* d_A;

CUDA_SAFE_CALL(cudaMalloc((void**) & d_A, mem_size_A));

float* d_B;

CUDA_SAFE_CALL(cudaMalloc((void**) & d_B, mem_size_B));

 

// копирование матриц на GPU

CUDA_SAFE_CALL(cudaMemcpy(d_A, h_A, mem_size_A, cudaMemcpyHostToDevice));

CUDA_SAFE_CALL(cudaMemcpy(d_B, h_B, mem_size_B, cudaMemcpyHostToDevice));

 

// выделение памяти на GPU для матрицы-результата

unsigned int size_C = WC * HC;

unsigned int mem_size_C = sizeof(float) * size_C;

float* d_C;

CUDA_SAFE_CALL(cudaMalloc((void**) & d_C, mem_size_C));

 

// создание и запуск таймера

unsigned int timer = 0;

CUT_SAFE_CALL(cutCreateTimer(& timer));

CUT_SAFE_CALL(cutStartTimer(timer));

 

// установка параметров выполнения ядра

dim3 threads(BLOCK_SIZE, BLOCK_SIZE);

dim3 grid(WC / threads.x, HC / threads.y);

 

// запуск ядра

matrixMul< < < grid, threads > > > (d_C, d_A, d_B, WA, WB);

 

// проверяем на наличие ошибок выполнения ядра

CUT_CHECK_ERROR(" Ошибка выполнения ядра");

 

// выделяем память на CPU для матрицы-результата

float* h_C = (float*) malloc(mem_size_C);

 

// копируем результат на CPU

CUDA_SAFE_CALL(cudaMemcpy(h_C, d_C, mem_size_C,

cudaMemcpyDeviceToHost));

 

// останавливаем таймер и освобождаем ресурсы

CUT_SAFE_CALL(cutStopTimer(timer));

printf(" Processing time: %f (ms) n", cutGetTimerValue(timer));

CUT_SAFE_CALL(cutDeleteTimer(timer));

 

// вычисляем произведение A*B на CPU для сравнения точности

float* reference = (float*) malloc(mem_size_C);

computeGold(reference, h_A, h_B, HA, WA, WB);

 

// проверяем результат

CUTBoolean res = cutCompareL2fe(reference, h_C, size_C, 1e-6f);

printf(" Тест %s n", (1 == res)? " ПРОЙДЕН": " ПРОВАЛЕН");

if (res! =1) printDiff(reference, h_C, WC, HC);

 

// освобождаем выделенную память

free(h_A);

free(h_B);

free(h_C);

free(reference);

CUDA_SAFE_CALL(cudaFree(d_A));

CUDA_SAFE_CALL(cudaFree(d_B));

CUDA_SAFE_CALL(cudaFree(d_C));

}

 

// Заполнение матрицы псевдослучайными числами.

void randomInit(float* data, int size)

{

for (int i = 0; i < size; ++i)

data[i] = rand() / (float)RAND_MAX;

}

 

void printDiff(float *data1, float *data2, int width, int height)

{

int i, j, k;

int error_count=0;

for (j=0; j< height; j++) {

for (i=0; i< width; i++) {

k = j*width+i;

if (data1[k]! = data2[k]) {

printf(" diff(%d, %d) CPU=%4.4f, GPU=%4.4f n", i, j, data1[k], data2[k]);

error_count++;

}

}

}

printf(" Количество ошибок = %d n", error_count);

}

 

// Файл matrixMul.h [68]

 

// Copyright 1993-2007 NVIDIA Corporation. All rights reserved.

 

#ifndef _MATRIXMUL_H_

#define _MATRIXMUL_H_

 

// Размер связки потоков

#define BLOCK_SIZE 16

 

// Размеры матрицы (для простоты кода выбраны кратными размеру связки потоков)

#define WA (3 * BLOCK_SIZE) // Matrix A width

#define HA (5 * BLOCK_SIZE) // Matrix A height

#define WB (8 * BLOCK_SIZE) // Matrix B width

#define HB WA // Matrix B height

#define WC WB // Matrix C width

#define HC HA // Matrix C height

 

#endif // _MATRIXMUL_H_

 

#ifndef _MATRIXMUL_H_

#define _MATRIXMUL_H_

 

// Размер связки потоков

#define BLOCK_SIZE 16

 

// Размеры матрицы (для простоты кода выбраны кратными размеру связки потоков)

#define WA (3 * BLOCK_SIZE) // Matrix A width

#define HA (5 * BLOCK_SIZE) // Matrix A height

#define WB (8 * BLOCK_SIZE) // Matrix B width

#define HB WA // Matrix B height

#define WC WB // Matrix C width

#define HC HA // Matrix C height

 

#endif // _MATRIXMUL_H_

 

// Файл matrixMul_gold.cpp [68]

 

// Copyright 1993-2007 NVIDIA Corporation. All rights reserved.

 

extern " C"

void computeGold(float*, const float*, const float*, unsigned int, unsigned int, unsigned int);

 

////////////////////////////////////////////////////////////////////////////////

// Вычисляем произведение матриц на CPU для проверки

// C = A * B

// @param C массив для хранения результата (память выделяется заранее)

// @param A матрица A

// @param B матрица B

// @param hA кол-во строк матрицы А

// @param wB кол-во столбцов матрицы B

////////////////////////////////////////////////////////////////////////////////

void

computeGold(float* C, const float* A, const float* B, unsigned int hA, unsigned int wA, unsigned int wB)

{

for (unsigned int i = 0; i < hA; ++i)

for (unsigned int j = 0; j < wB; ++j) {

double sum = 0;

for (unsigned int k = 0; k < wA; ++k) {

double a = A[i * wA + k];

double b = B[k * wB + j];

sum += a * b;

}

C[i * wB + j] = (float)sum;

}}

Волгоградский филиал