게임/오리진&스팀&유플레이2014. 1. 16. 23:36

간만에 지름!! 험블번들 X

다른건 그리 안끌리는데 Strike Suit Zero 라는거에 삘 꽃혀서 질러버렸다 ㅠㅠ
Surgeon Simulator 2013는 이미 있는데 해보지 않았고.. 지인에게나 줘버려야지 ㅋㅋㅋ



'게임 > 오리진&스팀&유플레이' 카테고리의 다른 글

MDK 할인!  (0) 2014.03.19
몰라 걍 질러  (0) 2014.03.10
간만에 게임 목록  (0) 2013.12.29
스팀 연말 할인 시작!!  (0) 2013.12.22
리눅스 스팀 설치!  (0) 2013.12.17
Posted by 구차니
개소리 왈왈2014. 1. 16. 23:19

아 힘드러 ㅠㅠ

아침에는 수영
오후에는 학원 
그리고 간간히 고용센터 방문


왜이리 힘들고 피곤할까 ㅠㅠ
백수인게 원래 더 바쁘고 힘들다지만 ㅠㅠ 

'개소리 왈왈' 카테고리의 다른 글

백수.. 5주.. 차?  (0) 2014.01.23
ㅋㅋㅋㅋㅋ 국민카드 ㅋㅋㅋㅋㅋㅋ 시발 ㅋㅋㅋㅋㅋ  (0) 2014.01.19
아이매직 BW-1430 블루투스/USB 겸용 풀 사이즈 키보드  (0) 2014.01.06
새해 복 많이 받으세요!  (4) 2014.01.01
수영강습을 신청하였다.  (0) 2013.12.31
Posted by 구차니
Programming/openCL & CUDA2014. 1. 16. 00:03

cuda 1차원 데이터, 2차원 처리 예제

Phenum2 945(3.0Ghz) / GTX650

cpu Time : 0.131000
gpu Time : 0.001000

일단.. 싱글 코어로 돌려서 130msec 정도 
GPU로 돌려서 동일 연산이 1msec 미만으로 걸린다.. ㄷㄷㄷ

일단 32*32*128*128*4 로
총 1024개의 쓰레드를 동시에 돌리고
16,384 개의 블럭으로 연산을 한다.
67,108,864 byte 대략 64MB + 64MB = 64MB 용량의 연산을 하는 소스이다.

#include < stdio.h >
#include < stdlib.h >
#include < time.h >

#include "cuda_runtime.h"

#define BLOCK_WID	128
#define THREAD_WID	32
#define ARRAY_SIZE	 (THREAD_WID * THREAD_WID * BLOCK_WID * BLOCK_WID)

__global__ void kernel_test(int *a, int *b, int *c)
{
	int idx = threadIdx.x +blockIdx.x * blockDim.x + (gridDim.x * blockDim.x) * (blockIdx.y * blockDim.y + threadIdx.y);
	c[idx] = a[idx] + b[idx];
}

void main()
{
	clock_t start_time, end_time;
	int *a, *b, *c, *res;
	int *dev_a,*dev_b,*dev_c;
	unsigned int idx = 0;
	dim3 block(BLOCK_WID,BLOCK_WID);
	dim3 thread(THREAD_WID,THREAD_WID);

	a = (int *)malloc(ARRAY_SIZE * sizeof(int));
	b = (int *)malloc(ARRAY_SIZE * sizeof(int));
	c = (int *)malloc(ARRAY_SIZE * sizeof(int));
	res = (int *)malloc(ARRAY_SIZE * sizeof(int));

	// initialize
	srand (time(NULL));
	for(idx = 0;idx < ARRAY_SIZE ; idx++)
	{
		a[idx] = rand() & 0xFFFF;
		b[idx] = rand() & 0xFFFF;
		c[idx] = 0;
	}

	start_time = clock();
	for(idx = 0;idx < ARRAY_SIZE ; idx++)
	{
		res[idx] = a[idx] + b[idx];
	}
	end_time = clock();
	printf("cpu Time : %f\n", ((double)(end_time-start_time)) / CLOCKS_PER_SEC); 

	cudaMalloc(&dev_a, ARRAY_SIZE * sizeof(int));
	cudaMalloc(&dev_b, ARRAY_SIZE * sizeof(int));
	cudaMalloc(&dev_c, ARRAY_SIZE * sizeof(int));

	cudaMemcpy(dev_a, a, ARRAY_SIZE * sizeof(int), cudaMemcpyHostToDevice);
	cudaMemcpy(dev_b, b, ARRAY_SIZE * sizeof(int), cudaMemcpyHostToDevice);

	start_time = clock();
	kernel_test<<<block,thread>>>(dev_a,dev_b,dev_c);
	end_time = clock();
	printf("gpu Time : %f\n", ((double)(end_time-start_time)) / CLOCKS_PER_SEC); 

	cudaMemcpy(c, dev_c, ARRAY_SIZE * sizeof(int), cudaMemcpyDeviceToHost);

	for(idx = 0;idx < ARRAY_SIZE ; idx++)
	{
		if(res[idx] != c[idx])
		{
			printf("%5d a:%5d b:%5d c:%5d", idx, a[idx], b[idx], c[idx]);
			if(res[idx] != c[idx])
					printf(" != ");
			else	printf(" == ");
			printf("r:%5d\n",res[idx]);
			break;
		}
	}

	cudaFree(dev_a);
	cudaFree(dev_b);
	cudaFree(dev_c);

	free(a);
	free(b);
	free(c);
	free(res);
}

'Programming > openCL & CUDA' 카테고리의 다른 글

cuda + openmp 적용 예제  (0) 2014.01.17
cuda + openmp 프로젝트 생성은....  (0) 2014.01.17
CUDA 쓰레드 계산  (0) 2014.01.15
visual studio 2008 nsight 실행..된거 맞나?  (0) 2014.01.15
nsight 설치... -_-  (0) 2014.01.15
Posted by 구차니
Programming/openCL & CUDA2014. 1. 15. 21:58

CUDA 쓰레드 계산

머리를 데굴데굴 굴려봐도 깔끔한 계산식이 안나온다..
걍 2차원 배열로 하는수 밖에 없으려나?

일단 1차원으로 할당한 녀석을 2차원 block/thread 로 구성시에 대한 수식이다.

threadIdx.x +blockIdx.x * blockDim.x
+ (gridDim.x * blockDim.x) * (blockIdx.y * blockDim.y + threadIdx.y);

근데.. 걍.. 데이터를 2차원 으로 해서 하는게 속 편할지도 모르겠네...


뻘짓 끝에 디버거로 돌려보니..
blockIdx 에는 thread block의 dimension이 들어간다....

다시 글들을 보니..
block에 대한 카운트와 인덱스는 gridDim 과 blockIdx에서
thread에 대한 카운트와 인덱스는 blockDim 과 threadIdx 에서 내장변수로 제공한다.

---
끄아아.. 벌써 세번째 글이었구나 ㅠㅠ

2011/01/16 - [Programming/openCL / CUDA] - CUDA 내장변수 - built in variable
2012/04/30 - [Programming/openCL / CUDA] - cuda 내장변수 

'Programming > openCL & CUDA' 카테고리의 다른 글

cuda + openmp 프로젝트 생성은....  (0) 2014.01.17
cuda 1차원 데이터, 2차원 처리 예제  (4) 2014.01.16
visual studio 2008 nsight 실행..된거 맞나?  (0) 2014.01.15
nsight 설치... -_-  (0) 2014.01.15
cuda 1 block 1차원 & 2차원 thread 예제  (0) 2014.01.14
Posted by 구차니
Programming/openCL & CUDA2014. 1. 15. 11:29

visual studio 2008 nsight 실행..된거 맞나?

Visual Studio 2008 SP1설치후
CUDA 5.5 재설치 - Nsight Visual Studio Edition
음.. 그런데 이녀석.. CUDA SDK에 포함되어서 굳이 Nvidia에서 가입/등록하고 안해도 되네?
물론.. SDK를 통채로 다운받아서 갈아 엎어야 하는 귀차니즘이 있긴 하겠지만..


Installed에 똭! 추가!


기본 값은 Launcher debugging enabled False이므로


True로 바꾸어 주고(위에것도 해야하려나?)


VS2008의 디버거를 사용하지 말고 그냥
Start CUDA Debugging을 누르면 지가 알아서 실행되고 프로파일링 한다.


머하다가 이게 떴더라..?


막 누르다 보니.. 정신없는데..
Application Control에 Launch를 누르면 프로그램이 실행되면서
Nsight에서 녹화하고 그 데이터를 기반으로 디버깅 정보를 출력한다.


Summary Report 아직 보는 법은 모르겠고


nsight를 설치해서인지 VS2008 자체 디버거로 실행시 kernel에서 내장 변수들이 출력된다.
근데.. 막 건너뛰는 기분..(마스터 쓰레드만 하나 출력되려나?)


'Programming > openCL & CUDA' 카테고리의 다른 글

cuda 1차원 데이터, 2차원 처리 예제  (4) 2014.01.16
CUDA 쓰레드 계산  (0) 2014.01.15
nsight 설치... -_-  (0) 2014.01.15
cuda 1 block 1차원 & 2차원 thread 예제  (0) 2014.01.14
cuda 5.5 조금은 더 빠르게 컴파일 하기  (0) 2014.01.14
Posted by 구차니
Programming/openCL & CUDA2014. 1. 15. 10:41

nsight 설치... -_-

Nsight는 VS2008일 경우 SP1 이상을 요구한다

Requirement not met: Service Pack 1 for Microsoft Visual Studio 2008 was not found.


하지만 문제는 Nsight 샘플 프로젝트가 없다는 것!!!


분위기로 봐서는.. CUDA 5.5 설치시에
VS2008 SP1이 없어서 Nsight for Visual Studio 2008이 설치 되지 않으면서 생긴 문제로 생각되니..
이래저래 CUDA 5.5도 전부 다시 재설치.. 끄아아 ㅠㅠ


[링크 : https://developer.nvidia.com/install-nsight-visual-studio-edition]

'Programming > openCL & CUDA' 카테고리의 다른 글

CUDA 쓰레드 계산  (0) 2014.01.15
visual studio 2008 nsight 실행..된거 맞나?  (0) 2014.01.15
cuda 1 block 1차원 & 2차원 thread 예제  (0) 2014.01.14
cuda 5.5 조금은 더 빠르게 컴파일 하기  (0) 2014.01.14
GTX650 / ion devicequery  (4) 2014.01.13
Posted by 구차니
Programming/openCL & CUDA2014. 1. 14. 19:53

cuda 1 block 1차원 & 2차원 thread 예제

visual studio 2008 / cuda 5.5 / GTX650 기준 작성

    0 a:26326 b:14567 c:40893 == r:40893
    1 a:20769 b:29469 c:50238 == r:50238
    2 a:19293 b:19828 c:39121 == r:39121
    3 a: 5720 b:16164 c:21884 == r:21884
    4 a:10116 b:16010 c:26126 == r:26126
    5 a:24503 b: 1380 c:25883 == r:25883
    6 a: 1261 b:20500 c:21761 == r:21761
    7 a:32527 b:14265 c:46792 == r:46792
    8 a: 6165 b: 1639 c: 7804 == r: 7804
    9 a:16881 b: 7619 c:24500 == r:24500
   10 a:14636 b: 3016 c:17652 == r:17652
   11 a:20766 b: 1675 c:22441 == r:22441
   12 a:24356 b: 3886 c:28242 == r:28242
   13 a: 9279 b:15721 c:25000 == r:25000
   14 a:20744 b:   74 c:20818 == r:20818
   15 a:24023 b:17957 c:41980 == r:41980
   16 a:  399 b:19653 c:20052 == r:20052
   17 a: 9077 b: 9308 c:18385 == r:18385
   18 a:18673 b:  713 c:19386 == r:19386
   19 a:17966 b:12837 c:30803 == r:30803
   20 a:28921 b:31938 c:60859 == r:60859
   21 a:20298 b:18933 c:39231 == r:39231
   22 a:18267 b:31334 c:49601 == r:49601
   23 a:17726 b:18368 c:36094 == r:36094
   24 a:10825 b:19187 c:30012 == r:30012
   25 a:15579 b: 9569 c:25148 == r:25148
   26 a:17217 b:27831 c:45048 == r:45048
   27 a: 2756 b:13884 c:16640 == r:16640
   28 a:25641 b:17878 c:43519 == r:43519
   29 a:10533 b:17954 c:28487 == r:28487
   30 a:15005 b:23112 c:38117 == r:38117
   31 a: 9634 b: 8053 c:17687 == r:17687 		 


#include < stdio.h >
#include < stdlib.h >
#include < time.h >

#include "cuda_runtime.h"

#define ARRAY_SIZE	32

__global__ void kernel_test(int *a, int *b, int *c)
{
	int idx = threadIdx.x;
	c[idx] = a[idx] + b[idx];
}

void main()
{
	int a[ARRAY_SIZE],b[ARRAY_SIZE],c[ARRAY_SIZE],res[ARRAY_SIZE];
	int *dev_a,*dev_b,*dev_c;
	int idx = 0;
	dim3 block(1);
	dim3 thread(ARRAY_SIZE);

	// initialize
	srand (time(NULL));
	for(idx = 0;idx < ARRAY_SIZE ; idx++)
	{
		a[idx] = rand();
		b[idx] = rand();
		res[idx] = a[idx] + b[idx];
		c[idx] = 0;
	}

	cudaMalloc(&dev_a, ARRAY_SIZE * sizeof(int));
	cudaMalloc(&dev_b, ARRAY_SIZE * sizeof(int));
	cudaMalloc(&dev_c, ARRAY_SIZE * sizeof(int));

	cudaMemcpy(dev_a, a, ARRAY_SIZE * sizeof(int), cudaMemcpyHostToDevice);
	cudaMemcpy(dev_b, b, ARRAY_SIZE * sizeof(int), cudaMemcpyHostToDevice);

	kernel_test<<<block,thread>>>(dev_a,dev_b,dev_c);

	cudaMemcpy(c, dev_c, ARRAY_SIZE * sizeof(int), cudaMemcpyDeviceToHost);

	for(idx = 0;idx < ARRAY_SIZE ; idx++)
	{
		printf("%5d a:%5d b:%5d c:%5d", idx, a[idx], b[idx], c[idx]);
		if(res[idx] != c[idx])
				printf(" != ");
		else	printf(" == ");
		printf("r:%5d\n",res[idx]);
	}

	cudaFree(dev_a);
	cudaFree(dev_b);
	cudaFree(dev_c);
}


#define ARRAY_SIZE	64

__global__ void kernel_test(int *a, int *b, int *c)
{
	int idx = threadIdx.x + threadIdx.y * 8;
	c[idx] = a[idx] + b[idx];
}

void main()
{
	dim3 block(1);
	dim3 thread(8,8);
}

'Programming > openCL & CUDA' 카테고리의 다른 글

visual studio 2008 nsight 실행..된거 맞나?  (0) 2014.01.15
nsight 설치... -_-  (0) 2014.01.15
cuda 5.5 조금은 더 빠르게 컴파일 하기  (0) 2014.01.14
GTX650 / ion devicequery  (4) 2014.01.13
cuda dim3 변수 초기화  (0) 2014.01.13
Posted by 구차니
Programming/openCL & CUDA2014. 1. 14. 17:14

cuda 5.5 조금은 더 빠르게 컴파일 하기

vs2008의 문제인지 cuda5.5의 변경점인지 모르겠지만
컴파일시 하나의 파일에 대해서 경고가 여러번 뜨는 것 봐서는
여러번의 컴파일을 시도하는 것으로 생각되어 프로젝트 옵션을 뒤져보고 대충 테스트 해본결과

기본값으로 GPU Architecture가 sm_10과 sm_20 두개가 설정이 되어 있어 2번의 컴파일을 수행하고 있었다.
그래픽 카드의 아키텍쳐를 확실히 안하면 제대로 설정해서 한번만 컴파일 하는 것도
컴파일 시간 단축에 도움이 많이 될 듯 하다.

1,2에 sm_10 / sm_20을

 
GTX650을 쓰고 있으니 sm_30으로 설정!

 

'Programming > openCL & CUDA' 카테고리의 다른 글

nsight 설치... -_-  (0) 2014.01.15
cuda 1 block 1차원 & 2차원 thread 예제  (0) 2014.01.14
GTX650 / ion devicequery  (4) 2014.01.13
cuda dim3 변수 초기화  (0) 2014.01.13
vs2008 cuda syntax highlight  (0) 2014.01.13
Posted by 구차니
Programming/openCL & CUDA2014. 1. 13. 22:12

GTX650 / ion devicequery

가장 중요한 차이점 아래 스샷 참조
왼쪽이 ion 오른쪽이 GTX650 이다.


전체 비교


사이즈 문제로 텍스트 버전은 아래에


'Programming > openCL & CUDA' 카테고리의 다른 글

cuda 1 block 1차원 & 2차원 thread 예제  (0) 2014.01.14
cuda 5.5 조금은 더 빠르게 컴파일 하기  (0) 2014.01.14
cuda dim3 변수 초기화  (0) 2014.01.13
vs2008 cuda syntax highlight  (0) 2014.01.13
VS2008 + cuda5.5 프로젝트 생성  (0) 2014.01.13
Posted by 구차니
Programming/openCL & CUDA2014. 1. 13. 17:58

cuda dim3 변수 초기화

dim3는 uint3 형(unsigned int)으로 x,y,z 3개의 변수를 지니는 구조체다.
c++ 스타일과 c 스타일로 초기화가 각각 가능하며

C++ 스타일로는
dim3 valname(x,y,z);
dim3 valname(x,y);
dim3 valname(x); 
식으로 값이 없는 건 1로 선언되며

C 스타일로는
dim3 valname = {x,y,z};
으로 선언된다.
---
엥? c 스타일로는 선언 안되는데?
dim3 testdim3 = {1,1,1};
dim3 testdim2 = {1,1}
dim3 testdim1 = {1} 

error: initialization with "{...}" is not allowed for object of type "dim3"
error: initialization with "{...}" is not allowed for object of type "dim3"
error: initialization with "{...}" is not allowed for object of type "dim3" 
---

C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v5.5\include\vector_types.h
struct __device_builtin__ dim3
{
    unsigned int x, y, z;
#if defined(__cplusplus)
    __host__ __device__ dim3(unsigned int vx = 1, unsigned int vy = 1, unsigned int vz = 1) : x(vx), y(vy), z(vz) {}
    __host__ __device__ dim3(uint3 v) : x(v.x), y(v.y), z(v.z) {}
    __host__ __device__ operator uint3(void) { uint3 t; t.x = x; t.y = y; t.z = z; return t; }
#endif /* __cplusplus */
};

typedef __device_builtin__ struct dim3 dim3;

[링크 : http://choorucode.com/2011/02/16/cuda-dim3/]

'Programming > openCL & CUDA' 카테고리의 다른 글

cuda 5.5 조금은 더 빠르게 컴파일 하기  (0) 2014.01.14
GTX650 / ion devicequery  (4) 2014.01.13
vs2008 cuda syntax highlight  (0) 2014.01.13
VS2008 + cuda5.5 프로젝트 생성  (0) 2014.01.13
nvidia ion + ubuntu 12.04 LTS + cuda5.5  (0) 2014.01.08
Posted by 구차니

티스토리툴바