#include <iostream>
#include <omp.h>

int main() {
    // 병렬 영역 시작!
    #pragma omp parallel
    {
        int thread_id = omp_get_thread_num();
        int total_threads = omp_get_num_threads();
        
        std::cout << "안녕! 나는 스레드 " << thread_id 
                  << " / 전체 " << total_threads << "개 중 하나야!" 
                  << std::endl;
    }
    
    return 0;
}

#include <iostream>
#include <vector>
#include <omp.h>
#include <chrono>

// 순차 버전
double sequential_sum(const std::vector<double>& data) {
    double sum = 0.0;
    for (size_t i = 0; i < data.size(); i++) {
        sum += data[i];
    }
    return sum;
}

// OpenMP 병렬 버전
double parallel_sum(const std::vector<double>& data) {
    double sum = 0.0;
    
    #pragma omp parallel for reduction(+:sum)
    for (size_t i = 0; i < data.size(); i++) {
        sum += data[i];
    }
    
    return sum;
}

int main() {
    const size_t N = 100000000; // 1억 개의 데이터
    std::vector<double> data(N, 1.0);
    
    // 순차 실행 시간 측정
    auto start = std::chrono::high_resolution_clock::now();
    double result1 = sequential_sum(data);
    auto end = std::chrono::high_resolution_clock::now();
    auto duration1 = std::chrono::duration_cast<std::chrono::milliseconds>(end - start);
    
    std::cout << "순차 실행 결과: " << result1 << std::endl;
    std::cout << "순차 실행 시간: " << duration1.count() << "ms" << std::endl;
    
    // 병렬 실행 시간 측정
    start = std::chrono::high_resolution_clock::now();
    double result2 = parallel_sum(data);
    end = std::chrono::high_resolution_clock::now();
    auto duration2 = std::chrono::duration_cast<std::chrono::milliseconds>(end - start);
    
    std::cout << "병렬 실행 결과: " << result2 << std::endl;
    std::cout << "병렬 실행 시간: " << duration2.count() << "ms" << std::endl;
    std::cout << "속도 향상: " << (double)duration1.count() / duration2.count() << "배!" << std::endl;
    
    return 0;
}

#include <vector>
#include <omp.h>

struct Pixel {
    unsigned char r, g, b;
};

class Image {
private:
    std::vector<Pixel> data;
    int width, height;
    
public:
    Image(int w, int h) : width(w), height(h), data(w * h) {}
    
    // 그레이스케일 변환 (병렬 버전)
    void convertToGrayscale() {
        #pragma omp parallel for collapse(2)
        for (int y = 0; y < height; y++) {
            for (int x = 0; x < width; x++) {
                int idx = y * width + x;
                Pixel& p = data[idx];
                
                // 그레이스케일 공식: 0.299R + 0.587G + 0.114B
                unsigned char gray = static_cast<unsigned char>(
                    0.299 * p.r + 0.587 * p.g + 0.114 * p.b
                );
                
                p.r = p.g = p.b = gray;
            }
        }
    }
    
    // 블러 효과 (병렬 버전)
    void applyBlur(int radius) {
        std::vector<Pixel> temp = data;
        
        #pragma omp parallel for collapse(2)
        for (int y = radius; y < height - radius; y++) {
            for (int x = radius; x < width - radius; x++) {
                int sumR = 0, sumG = 0, sumB = 0;
                int count = 0;
                
                // 주변 픽셀 평균 계산
                for (int dy = -radius; dy <= radius; dy++) {
                    for (int dx = -radius; dx <= radius; dx++) {
                        int idx = (y + dy) * width + (x + dx);
                        sumR += temp[idx].r;
                        sumG += temp[idx].g;
                        sumB += temp[idx].b;
                        count++;
                    }
                }
                
                int idx = y * width + x;
                data[idx].r = sumR / count;
                data[idx].g = sumG / count;
                data[idx].b = sumB / count;
            }
        }
    }
};

#include <iostream>
#include <cuda_runtime.h>

// __global__은 이 함수가 GPU에서 실행됨을 의미해
__global__ void helloFromGPU() {
    int threadId = threadIdx.x + blockIdx.x * blockDim.x;
    printf("안녕! 나는 GPU 스레드 %d야!\n", threadId);
}

int main() {
    std::cout << "CPU에서 시작!" << std::endl;
    
    // 10개의 블록, 각 블록당 10개의 스레드 = 총 100개 스레드
    helloFromGPU<<<10, 10>>>();
    
    // GPU 작업이 끝날 때까지 대기
    cudaDeviceSynchronize();
    
    std::cout << "CPU에서 종료!" << std::endl;
    return 0;
}

#include <iostream>
#include <cuda_runtime.h>
#include <chrono>

// CPU 버전
void vectorAddCPU(const float* a, const float* b, float* c, int n) {
    for (int i = 0; i < n; i++) {
        c[i] = a[i] + b[i];
    }
}

// GPU 커널
__global__ void vectorAddGPU(const float* a, const float* b, float* c, int n) {
    int idx = threadIdx.x + blockIdx.x * blockDim.x;
    
    if (idx < n) {
        c[idx] = a[idx] + b[idx];
    }
}

int main() {
    const int N = 10000000; // 1천만 개의 원소
    const int bytes = N * sizeof(float);
    
    // 호스트 메모리 할당
    float *h_a = new float[N];
    float *h_b = new float[N];
    float *h_c = new float[N];
    
    // 데이터 초기화
    for (int i = 0; i < N; i++) {
        h_a[i] = static_cast<float>(i);
        h_b[i] = static_cast<float>(i * 2);
    }
    
    // CPU 버전 실행 시간 측정
    auto start = std::chrono::high_resolution_clock::now();
    vectorAddCPU(h_a, h_b, h_c, N);
    auto end = std::chrono::high_resolution_clock::now();
    auto cpuTime = std::chrono::duration_cast<std::chrono::milliseconds>(end - start);
    
    std::cout << "CPU 실행 시간: " << cpuTime.count() << "ms" << std::endl;
    
    // 디바이스 메모리 할당
    float *d_a, *d_b, *d_c;
    cudaMalloc(&d_a, bytes);
    cudaMalloc(&d_b, bytes);
    cudaMalloc(&d_c, bytes);
    
    // GPU 버전 실행 시간 측정 (메모리 복사 포함)
    start = std::chrono::high_resolution_clock::now();
    
    // 호스트 → 디바이스 메모리 복사
    cudaMemcpy(d_a, h_a, bytes, cudaMemcpyHostToDevice);
    cudaMemcpy(d_b, h_b, bytes, cudaMemcpyHostToDevice);
    
    // 커널 실행 설정
    int threadsPerBlock = 256;
    int blocksPerGrid = (N + threadsPerBlock - 1) / threadsPerBlock;
    
    vectorAddGPU<<<blocksPerGrid, threadsPerBlock>>>(d_a, d_b, d_c, N);
    
    // 디바이스 → 호스트 메모리 복사
    cudaMemcpy(h_c, d_c, bytes, cudaMemcpyDeviceToHost);
    
    end = std::chrono::high_resolution_clock::now();
    auto gpuTime = std::chrono::duration_cast<std::chrono::milliseconds>(end - start);
    
    std::cout << "GPU 실행 시간 (메모리 복사 포함): " << gpuTime.count() << "ms" << std::endl;
    std::cout << "속도 향상: " << (double)cpuTime.count() / gpuTime.count() << "배!" << std::endl;
    
    // 결과 검증
    bool correct = true;
    for (int i = 0; i < N; i++) {
        if (abs(h_c[i] - (h_a[i] + h_b[i])) > 1e-5) {
            correct = false;
            break;
        }
    }
    std::cout << "결과 검증: " << (correct ? "성공! ✅" : "실패! ❌") << std::endl;
    
    // 메모리 해제
    delete[] h_a;
    delete[] h_b;
    delete[] h_c;
    cudaFree(d_a);
    cudaFree(d_b);
    cudaFree(d_c);
    
    return 0;
}

#include <cuda_runtime.h>

#define TILE_SIZE 16

// 기본 버전 (느림)
__global__ void matrixMulBasic(float* A, float* B, float* C, int N) {
    int row = blockIdx.y * blockDim.y + threadIdx.y;
    int col = blockIdx.x * blockDim.x + threadIdx.x;
    
    if (row < N && col < N) {
        float sum = 0.0f;
        for (int k = 0; k < N; k++) {
            sum += A[row * N + k] * B[k * N + col];
        }
        C[row * N + col] = sum;
    }
}

// 최적화 버전 (빠름!) - Shared Memory 사용
__global__ void matrixMulOptimized(float* A, float* B, float* C, int N) {
    // Shared Memory 선언
    __shared__ float tileA[TILE_SIZE][TILE_SIZE];
    __shared__ float tileB[TILE_SIZE][TILE_SIZE];
    
    int row = blockIdx.y * TILE_SIZE + threadIdx.y;
    int col = blockIdx.x * TILE_SIZE + threadIdx.x;
    
    float sum = 0.0f;
    
    // 타일 단위로 처리
    for (int t = 0; t < (N + TILE_SIZE - 1) / TILE_SIZE; t++) {
        // Global Memory → Shared Memory로 데이터 로드
        if (row < N && t * TILE_SIZE + threadIdx.x < N) {
            tileA[threadIdx.y][threadIdx.x] = A[row * N + t * TILE_SIZE + threadIdx.x];
        } else {
            tileA[threadIdx.y][threadIdx.x] = 0.0f;
        }
        
        if (col < N && t * TILE_SIZE + threadIdx.y < N) {
            tileB[threadIdx.y][threadIdx.x] = B[(t * TILE_SIZE + threadIdx.y) * N + col];
        } else {
            tileB[threadIdx.y][threadIdx.x] = 0.0f;
        }
        
        // 모든 스레드가 데이터 로드를 완료할 때까지 대기
        __syncthreads();
        
        // Shared Memory에서 계산 (빠름!)
        for (int k = 0; k < TILE_SIZE; k++) {
            sum += tileA[threadIdx.y][k] * tileB[k][threadIdx.x];
        }
        
        // 다음 타일로 넘어가기 전 동기화
        __syncthreads();
    }
    
    if (row < N && col < N) {
        C[row * N + col] = sum;
    }
}

#include <iostream>
#include <omp.h>
#include <cuda_runtime.h>

// GPU 커널
__global__ void processChunkGPU(float* data, int size) {
    int idx = threadIdx.x + blockIdx.x * blockDim.x;
    if (idx < size) {
        // GPU에서 처리할 작업
        data[idx] = data[idx] * data[idx];
    }
}

// CPU 함수
void processChunkCPU(float* data, int size) {
    #pragma omp parallel for
    for (int i = 0; i < size; i++) {
        // CPU에서 처리할 작업
        data[i] = data[i] * data[i];
    }
}

int main() {
    const int TOTAL_SIZE = 10000000;
    const int GPU_PORTION = 8000000; // 80%는 GPU에서
    const int CPU_PORTION = 2000000; // 20%는 CPU에서
    
    float *data = new float[TOTAL_SIZE];
    float *d_data;
    
    // 데이터 초기화
    for (int i = 0; i < TOTAL_SIZE; i++) {
        data[i] = static_cast<float>(i);
    }
    
    // GPU 메모리 할당
    cudaMalloc(&d_data, GPU_PORTION * sizeof(float));
    
    // CPU와 GPU 동시 실행!
    #pragma omp parallel sections
    {
        #pragma omp section
        {
            // GPU 작업
            cudaMemcpy(d_data, data, GPU_PORTION * sizeof(float), 
                      cudaMemcpyHostToDevice);
            
            int threadsPerBlock = 256;
            int blocksPerGrid = (GPU_PORTION + threadsPerBlock - 1) / threadsPerBlock;
            processChunkGPU<<<blocksPerGrid, threadsPerBlock>>>(d_data, GPU_PORTION);
            
            cudaMemcpy(data, d_data, GPU_PORTION * sizeof(float), 
                      cudaMemcpyDeviceToHost);
        }
        
        #pragma omp section
        {
            // CPU 작업
            processChunkCPU(data + GPU_PORTION, CPU_PORTION);
        }
    }
    
    std::cout << "CPU와 GPU가 함께 작업 완료! 🎉" << std::endl;
    
    // 정리
    delete[] data;
    cudaFree(d_data);
    
    return 0;
}

#include <iostream>
#include <vector>
#include <omp.h>
#include <cuda_runtime.h>

// 픽셀 구조체
struct Pixel {
    unsigned char r, g, b, a;
};

// CUDA 커널: 가우시안 블러
__global__ void gaussianBlurKernel(Pixel* input, Pixel* output, 
                                   int width, int height, float sigma) {
    int x = blockIdx.x * blockDim.x + threadIdx.x;
    int y = blockIdx.y * blockDim.y + threadIdx.y;
    
    if (x >= width || y >= height) return;
    
    const int radius = 5;
    float sumR = 0, sumG = 0, sumB = 0, sumWeight = 0;
    
    for (int dy = -radius; dy <= radius; dy++) {
        for (int dx = -radius; dx <= radius; dx++) {
            int nx = x + dx;
            int ny = y + dy;
            
            if (nx >= 0 && nx < width && ny >= 0 && ny < height) {
                float distance = sqrtf(dx * dx + dy * dy);
                float weight = expf(-(distance * distance) / (2 * sigma * sigma));
                
                int idx = ny * width + nx;
                sumR += input[idx].r * weight;
                sumG += input[idx].g * weight;
                sumB += input[idx].b * weight;
                sumWeight += weight;
            }
        }
    }
    
    int idx = y * width + x;
    output[idx].r = static_cast<unsigned char>(sumR / sumWeight);
    output[idx].g = static_cast<unsigned char>(sumG / sumWeight);
    output[idx].b = static_cast<unsigned char>(sumB / sumWeight);
    output[idx].a = input[idx].a;
}

// CUDA 커널: 엣지 검출 (Sobel)
__global__ void sobelEdgeKernel(Pixel* input, Pixel* output, 
                                int width, int height) {
    int x = blockIdx.x * blockDim.x + threadIdx.x;
    int y = blockIdx.y * blockDim.y + threadIdx.y;
    
    if (x >= width || y >= height || x == 0 || y == 0 || 
        x == width - 1 || y == height - 1) return;
    
    // Sobel 연산자
    int gx = 0, gy = 0;
    
    for (int dy = -1; dy <= 1; dy++) {
        for (int dx = -1; dx <= 1; dx++) {
            int idx = (y + dy) * width + (x + dx);
            int gray = (input[idx].r + input[idx].g + input[idx].b) / 3;
            
            // Gx 계산
            if (dx == -1) gx -= gray * (dy == 0 ? 2 : 1);
            if (dx == 1) gx += gray * (dy == 0 ? 2 : 1);
            
            // Gy 계산
            if (dy == -1) gy -= gray * (dx == 0 ? 2 : 1);
            if (dy == 1) gy += gray * (dx == 0 ? 2 : 1);
        }
    }
    
    int magnitude = static_cast<int>(sqrtf(gx * gx + gy * gy));
    magnitude = min(255, magnitude);
    
    int idx = y * width + x;
    output[idx].r = output[idx].g = output[idx].b = magnitude;
    output[idx].a = 255;
}

class ImageProcessor {
private:
    std::vector<Pixel> hostImage;
    Pixel *deviceInput, *deviceOutput;
    int width, height;
    
public:
    ImageProcessor(int w, int h) : width(w), height(h) {
        hostImage.resize(w * h);
        
        size_t bytes = w * h * sizeof(Pixel);
        cudaMalloc(&deviceInput, bytes);
        cudaMalloc(&deviceOutput, bytes);
    }
    
    ~ImageProcessor() {
        cudaFree(deviceInput);
        cudaFree(deviceOutput);
    }
    
    // OpenMP로 이미지 생성 (테스트용)
    void generateTestImage() {
        #pragma omp parallel for collapse(2)
        for (int y = 0; y < height; y++) {
            for (int x = 0; x < width; x++) {
                int idx = y * width + x;
                
                // 그라디언트 패턴
                hostImage[idx].r = (x * 255) / width;
                hostImage[idx].g = (y * 255) / height;
                hostImage[idx].b = ((x + y) * 255) / (width + height);
                hostImage[idx].a = 255;
            }
        }
        std::cout << "OpenMP로 테스트 이미지 생성 완료! ✅" << std::endl;
    }
    
    // CUDA로 가우시안 블러 적용
    void applyGaussianBlur(float sigma) {
        size_t bytes = width * height * sizeof(Pixel);
        
        // 호스트 → 디바이스
        cudaMemcpy(deviceInput, hostImage.data(), bytes, cudaMemcpyHostToDevice);
        
        // 커널 실행
        dim3 blockSize(16, 16);
        dim3 gridSize((width + 15) / 16, (height + 15) / 16);
        
        gaussianBlurKernel<<<gridSize, blockSize>>>(
            deviceInput, deviceOutput, width, height, sigma);
        
        // 디바이스 → 호스트
        cudaMemcpy(hostImage.data(), deviceOutput, bytes, cudaMemcpyDeviceToHost);
        
        std::cout << "CUDA로 가우시안 블러 적용 완료! ✅" << std::endl;
    }
    
    // CUDA로 엣지 검출
    void applyEdgeDetection() {
        size_t bytes = width * height * sizeof(Pixel);
        
        cudaMemcpy(deviceInput, hostImage.data(), bytes, cudaMemcpyHostToDevice);
        
        dim3 blockSize(16, 16);
        dim3 gridSize((width + 15) / 16, (height + 15) / 16);
        
        sobelEdgeKernel<<<gridSize, blockSize>>>(
            deviceInput, deviceOutput, width, height);
        
        cudaMemcpy(hostImage.data(), deviceOutput, bytes, cudaMemcpyDeviceToHost);
        
        std::cout << "CUDA로 엣지 검출 완료! ✅" << std::endl;
    }
    
    // OpenMP로 밝기 조정
    void adjustBrightness(float factor) {
        #pragma omp parallel for
        for (int i = 0; i < width * height; i++) {
            hostImage[i].r = min(255, static_cast<int>(hostImage[i].r * factor));
            hostImage[i].g = min(255, static_cast<int>(hostImage[i].g * factor));
            hostImage[i].b = min(255, static_cast<int>(hostImage[i].b * factor));
        }
        std::cout << "OpenMP로 밝기 조정 완료! ✅" << std::endl;
    }
    
    // OpenMP로 통계 계산
    void calculateStatistics() {
        long long sumR = 0, sumG = 0, sumB = 0;
        
        #pragma omp parallel for reduction(+:sumR,sumG,sumB)
        for (int i = 0; i < width * height; i++) {
            sumR += hostImage[i].r;
            sumG += hostImage[i].g;
            sumB += hostImage[i].b;
        }
        
        int totalPixels = width * height;
        std::cout << "평균 색상 - R: " << (sumR / totalPixels) 
                  << ", G: " << (sumG / totalPixels)
                  << ", B: " << (sumB / totalPixels) << std::endl;
    }
};

int main() {
    std::cout << "🎨 고성능 이미지 처리 시스템 시작!" << std::endl;
    std::cout << "========================================" << std::endl;
    
    // 4K 해상도 이미지
    const int WIDTH = 3840;
    const int HEIGHT = 2160;
    
    ImageProcessor processor(WIDTH, HEIGHT);
    
    // 1단계: OpenMP로 이미지 생성
    std::cout << "\n[1단계] 이미지 생성 중..." << std::endl;
    processor.generateTestImage();
    
    // 2단계: OpenMP로 전처리
    std::cout << "\n[2단계] 전처리 중..." << std::endl;
    processor.adjustBrightness(1.2f);
    
    // 3단계: CUDA로 필터 적용
    std::cout << "\n[3단계] GPU 필터 적용 중..." << std::endl;
    processor.applyGaussianBlur(2.0f);
    
    // 4단계: CUDA로 엣지 검출
    std::cout << "\n[4단계] 엣지 검출 중..." << std::endl;
    processor.applyEdgeDetection();
    
    // 5단계: OpenMP로 통계 계산
    std::cout << "\n[5단계] 통계 계산 중..." << std::endl;
    processor.calculateStatistics();
    
    std::cout << "\n========================================" << std::endl;
    std::cout << "✨ 모든 처리 완료! 4K 이미지를 실시간으로 처리했어!" << std::endl;
    
    return 0;
}

// static: 작업을 균등하게 미리 분배 (빠르지만 불균형 가능)
#pragma omp parallel for schedule(static)
for (int i = 0; i < N; i++) {
    // 작업 시간이 비슷할 때 좋음
}

// dynamic: 작업을 동적으로 분배 (균형 좋지만 오버헤드 있음)
#pragma omp parallel for schedule(dynamic, 10)
for (int i = 0; i < N; i++) {
    // 작업 시간이 다를 때 좋음
}

// guided: 처음엔 큰 청크, 나중엔 작은 청크
#pragma omp parallel for schedule(guided)
for (int i = 0; i < N; i++) {
    // 작업 시간이 점점 줄어들 때 좋음
}

// auto: 컴파일러가 알아서 선택
#pragma omp parallel for schedule(auto)
for (int i = 0; i < N; i++) {
    // 잘 모르겠으면 이거!
}

// 나쁜 예: False Sharing 발생
struct BadCounter {
    int count[8]; // 모두 같은 캐시 라인에 있을 수 있음
};

BadCounter counter;
#pragma omp parallel for
for (int i = 0; i < 8; i++) {
    for (int j = 0; j < 1000000; j++) {
        counter.count[i]++; // 캐시 라인 경쟁!
    }
}

// 좋은 예: 패딩으로 캐시 라인 분리
struct GoodCounter {
    alignas(64) int count; // 64바이트 정렬 (캐시 라인 크기)
};

GoodCounter counters[8];
#pragma omp parallel for
for (int i = 0; i < 8; i++) {
    for (int j = 0; j < 1000000; j++) {
        counters[i].count++; // 각자 다른 캐시 라인!
    }
}

// 나쁜 예: 비연속적 접근
__global__ void badMemoryAccess(float* data, int width, int height) {
    int x = blockIdx.x * blockDim.x + threadIdx.x;
    int y = blockIdx.y * blockDim.y + threadIdx.y;
    
    // 열 우선 접근 - 느림!
    int idx = x * height + y;
    data[idx] = data[idx] * 2.0f;
}

// 좋은 예: 연속적 접근
__global__ void goodMemoryAccess(float* data, int width, int height) {
    int x = blockIdx.x * blockDim.x + threadIdx.x;
    int y = blockIdx.y * blockDim.y + threadIdx.y;
    
    // 행 우선 접근 - 빠름!
    int idx = y * width + x;
    data[idx] = data[idx] * 2.0f;
}

#include <cuda_runtime.h>

void processWithStreams(float* h_data, int totalSize) {
    const int NUM_STREAMS = 4;
    const int chunkSize = totalSize / NUM_STREAMS;
    
    cudaStream_t streams[NUM_STREAMS];
    float *d_data[NUM_STREAMS];
    
    // 스트림 생성
    for (int i = 0; i < NUM_STREAMS; i++) {
        cudaStreamCreate(&streams[i]);
        cudaMalloc(&d_data[i], chunkSize * sizeof(float));
    }
    
    // 비동기 실행 - 모든 스트림이 동시에 작동!
    for (int i = 0; i < NUM_STREAMS; i++) {
        int offset = i * chunkSize;
        
        // 비동기 메모리 복사
        cudaMemcpyAsync(d_data[i], h_data + offset, 
                       chunkSize * sizeof(float),
                       cudaMemcpyHostToDevice, streams[i]);
        
        // 비동기 커널 실행
        int threadsPerBlock = 256;
        int blocksPerGrid = (chunkSize + threadsPerBlock - 1) / threadsPerBlock;
        processKernel<<<blocksPerGrid, threadsPerBlock, 0, streams[i]>>>(
            d_data[i], chunkSize);
        
        // 비동기 메모리 복사
        cudaMemcpyAsync(h_data + offset, d_data[i],
                       chunkSize * sizeof(float),
                       cudaMemcpyDeviceToHost, streams[i]);
    }
    
    // 모든 스트림 완료 대기
    for (int i = 0; i < NUM_STREAMS; i++) {
        cudaStreamSynchronize(streams[i]);
        cudaStreamDestroy(streams[i]);
        cudaFree(d_data[i]);
    }
}

#include <iostream>
#include <chrono>
#include <cuda_runtime.h>

class PerformanceTimer {
private:
    std::chrono::high_resolution_clock::time_point start;
    std::string name;
    
public:
    PerformanceTimer(const std::string& n) : name(n) {
        start = std::chrono::high_resolution_clock::now();
    }
    
    ~PerformanceTimer() {
        auto end = std::chrono::high_resolution_clock::now();
        auto duration = std::chrono::duration_cast<std::chrono::microseconds>(end - start);
        std::cout << name << ": " << duration.count() / 1000.0 << "ms" << std::endl;
    }
};

class CUDATimer {
private:
    cudaEvent_t start, stop;
    std::string name;
    
public:
    CUDATimer(const std::string& n) : name(n) {
        cudaEventCreate(&start);
        cudaEventCreate(&stop);
        cudaEventRecord(start);
    }
    
    ~CUDATimer() {
        cudaEventRecord(stop);
        cudaEventSynchronize(stop);
        
        float milliseconds = 0;
        cudaEventElapsedTime(&milliseconds, start, stop);
        std::cout << name << ": " << milliseconds << "ms" << std::endl;
        
        cudaEventDestroy(start);
        cudaEventDestroy(stop);
    }
};

// 사용 예시
void benchmarkExample() {
    const int N = 10000000;
    float *data = new float[N];
    
    {
        PerformanceTimer timer("CPU 초기화");
        #pragma omp parallel for
        for (int i = 0; i < N; i++) {
            data[i] = static_cast<float>(i);
        }
    }
    
    float *d_data;
    cudaMalloc(&d_data, N * sizeof(float));
    
    {
        CUDATimer timer("GPU 메모리 복사");
        cudaMemcpy(d_data, data, N * sizeof(float), cudaMemcpyHostToDevice);
    }
    
    {
        CUDATimer timer("GPU 커널 실행");
        int threadsPerBlock = 256;
        int blocksPerGrid = (N + threadsPerBlock - 1) / threadsPerBlock;
        processKernel<<<blocksPerGrid, threadsPerBlock>>>(d_data, N);
    }
    
    delete[] data;
    cudaFree(d_data);
}

// 간단한 파티클 시스템
struct Particle {
    float x, y, z;      // 위치
    float vx, vy, vz;   // 속도
    float mass;         // 질량
};

__global__ void updateParticles(Particle* particles, int count, float dt) {
    int idx = blockIdx.x * blockDim.x + threadIdx.x;
    if (idx >= count) return;
    
    Particle& p = particles[idx];
    
    // 중력 적용
    p.vy -= 9.8f * dt;
    
    // 위치 업데이트
    p.x += p.vx * dt;
    p.y += p.vy * dt;
    p.z += p.vz * dt;
    
    // 바닥 충돌
    if (p.y < 0.0f) {
        p.y = 0.0f;
        p.vy = -p.vy * 0.8f; // 반발 계수
    }
}

// 수만 개의 파티클을 실시간으로 처리!
void simulateParticles(Particle* d_particles, int count) {
    int threadsPerBlock = 256;
    int blocksPerGrid = (count + threadsPerBlock - 1) / threadsPerBlock;
    
    updateParticles<<<blocksPerGrid, threadsPerBlock>>>(
        d_particles, count, 0.016f); // 60 FPS
}

__global__ void calculateForces(float3* positions, float3* forces, 
                               int numAtoms) {
    int i = blockIdx.x * blockDim.x + threadIdx.x;
    if (i >= numAtoms) return;
    
    float3 force = make_float3(0.0f, 0.0f, 0.0f);
    float3 pos_i = positions[i];
    
    // 모든 다른 원자와의 상호작용 계산
    for (int j = 0; j < numAtoms; j++) {
        if (i == j) continue;
        
        float3 pos_j = positions[j];
        float3 r = make_float3(
            pos_j.x - pos_i.x,
            pos_j.y - pos_i.y,
            pos_j.z - pos_i.z
        );
        
        float dist = sqrtf(r.x*r.x + r.y*r.y + r.z*r.z);
        
        if (dist < 10.0f && dist > 0.1f) {
            // Lennard-Jones 포텐셜
            float sigma = 1.0f;
            float epsilon = 1.0f;
            float s6 = powf(sigma / dist, 6);
            float s12 = s6 * s6;
            
            float f_magnitude = 24.0f * epsilon * (2.0f * s12 - s6) / dist;
            
            force.x += f_magnitude * r.x / dist;
            force.y += f_magnitude * r.y / dist;
            force.z += f_magnitude * r.z / dist;
        }
    }
    
    forces[i] = force;
}

__global__ void monteCarloOption(float* results, float S0, float K, 
                                 float r, float sigma, float T, 
                                 int numPaths, unsigned int seed) {
    int idx = blockIdx.x * blockDim.x + threadIdx.x;
    if (idx >= numPaths) return;
    
    // 각 스레드마다 다른 난수 시드
    curandState state;
    curand_init(seed, idx, 0, &state);
    
    // 주가 경로 시뮬레이션
    float S = S0;
    int steps = 252; // 1년 = 252 거래일
    float dt = T / steps;
    
    for (int i = 0; i < steps; i++) {
        float z = curand_normal(&state);
        S *= expf((r - 0.5f * sigma * sigma) * dt + sigma * sqrtf(dt) * z);
    }
    
    // 콜 옵션 페이오프
    results[idx] = fmaxf(S - K, 0.0f) * expf(-r * T);
}

// 수백만 개의 시나리오를 몇 초 만에 계산!

// 잘못된 코드
int sum = 0;
#pragma omp parallel for
for (int i = 0; i < 1000; i++) {
    sum += i; // 레이스 컨디션! 결과가 매번 달라져
}

// 올바른 코드 1: reduction 사용
int sum = 0;
#pragma omp parallel for reduction(+:sum)
for (int i = 0; i < 1000; i++) {
    sum += i; // 안전!
}

// 올바른 코드 2: critical 사용 (느림)
int sum = 0;
#pragma omp parallel for
for (int i = 0; i < 1000; i++) {
    #pragma omp critical
    {
        sum += i; // 안전하지만 느림
    }
}

// 올바른 코드 3: atomic 사용 (빠름)
int sum = 0;
#pragma omp parallel for
for (int i = 0; i < 1000; i++) {
    #pragma omp atomic
    sum += i; // 안전하고 빠름
}

// 잘못된 코드 - 데이터 의존성 있음
#pragma omp parallel for
for (int i = 1; i < N; i++) {
    array[i] = array[i-1] + 1; // i번째가 i-1번째에 의존!
}

// 이런 경우는 병렬화할 수 없어. 순차적으로 실행해야 해.

// 잘못된 코드
void badFunction() {
    float *d_data;
    cudaMalloc(&d_data, 1000 * sizeof(float));
    // ... 작업 ...
    // cudaFree를 안 함! 메모리 누수!
}

// 올바른 코드 - RAII 패턴 사용
class CUDAMemory {
private:
    void* ptr;
    size_t size;
    
public:
    CUDAMemory(size_t s) : size(s) {
        cudaMalloc(&ptr, size);
    }
    
    ~CUDAMemory() {
        cudaFree(ptr); // 자동으로 해제!
    }
    
    void* get() { return ptr; }
};

void goodFunction() {
    CUDAMemory mem(1000 * sizeof(float));
    // ... 작업 ...
    // 자동으로 해제됨!
}

// 잘못된 코드
myKernel<<<blocks, threads>>>(data);
// 오류 확인 안 함!

// 올바른 코드
myKernel<<<blocks, threads>>>(data);

// 커널 실행 오류 확인
cudaError_t err = cudaGetLastError();
if (err != cudaSuccess) {
    std::cerr << "커널 실행 오류: " << cudaGetErrorString(err) << std::endl;
}

// 커널 완료 대기 및 오류 확인
err = cudaDeviceSynchronize();
if (err != cudaSuccess) {
    std::cerr << "커널 동기화 오류: " << cudaGetErrorString(err) << std::endl;
}

// 잘못된 코드
__global__ void badKernel(float* data, int N) {
    int idx = blockIdx.x * blockDim.x + threadIdx.x;
    data[idx] = idx; // N을 초과할 수 있음!
}

// 올바른 코드
__global__ void goodKernel(float* data, int N) {
    int idx = blockIdx.x * blockDim.x + threadIdx.x;
    if (idx < N) { // 범위 체크!
        data[idx] = idx;
    }
}