valentinsavenko · August 1, 2025 17:15
diff --git a/amd_gfx1030_llama.md b/amd_gfx1030_llama.md
diff --git a/stress_hip.cpp b/stress_hip.cpp
 #include <hip/hip_runtime.h>
 #include <stdio.h>
 #include <stdlib.h>
 #include <time.h>

 int HIP_DEVICE_ID = 0; // Set to the desired device ID

 #define CHECK_HIP(cmd) { \
    hipError_t err = cmd; \
    if (err != hipSuccess) { \
        fprintf(stderr, "HIP Error: %s at %s:%d\n", hipGetErrorString(err), __FILE__, __LINE__); \
        exit(1); \
    } \
 }



 __global__ void matrixMulHeavy(const float *A, const float *B, float *C, int M, int N, int K) {
    int row = blockIdx.y * blockDim.y + threadIdx.y;
    int col = blockIdx.x * blockDim.x + threadIdx.x;
    if (row < M && col < N) {
        float sum = 0.0f;
        for (int i = 0; i < K; i++) {
            sum += A[row * K + i] * B[i * N + col];
        }
        // Add heavy compute to maximize GPU utilization
        for (int j = 0; j < 500; j++) { // Increased iterations
            sum = sum * 1.00001f + sinf(sum) * 0.0001f; // Math-heavy operations
        }
        C[row * N + col] = sum;
    }
 }

 int main() {
    // Matrix dimensions: A (M x K), B (K x N), C (M x N)
    int M = 23000;
    int N = 23000;
    int K = 23000;
    size_t size_A = (size_t)M * K * sizeof(float);
    size_t size_B = (size_t)K * N * sizeof(float);
    size_t size_C = (size_t)M * N * sizeof(float);

    printf("Allocating ~%.2f GB VRAM\n", (size_A + size_B + size_C) / (1024.0 * 1024.0 * 1024.0));

    // Set device to Radeon 6700S 
    CHECK_HIP(hipSetDevice(HIP_DEVICE_ID));
    hipDeviceProp_t props;
    CHECK_HIP(hipGetDeviceProperties(&props, HIP_DEVICE_ID));
    printf("Running on device: %s\n", props.name);

    // Allocate host memory
    float *h_A = (float*)malloc(size_A);
    float *h_B = (float*)malloc(size_B);
    float *h_C = (float*)malloc(size_C);
    if (!h_A || !h_B || !h_C) {
        fprintf(stderr, "Host memory allocation failed\n");
        exit(1);
    }

    // Initialize matrices
    srand(time(NULL));
    for (size_t i = 0; i < (size_t)M * K; i++) h_A[i] = rand() / (float)RAND_MAX;
    for (size_t i = 0; i < (size_t)K * N; i++) h_B[i] = rand() / (float)RAND_MAX;

    // Allocate device memory
    float *d_A, *d_B, *d_C;
    CHECK_HIP(hipMalloc(&d_A, size_A));
    CHECK_HIP(hipMalloc(&d_B, size_B));
    CHECK_HIP(hipMalloc(&d_C, size_C));

    // Copy inputs to device
    CHECK_HIP(hipMemcpy(d_A, h_A, size_A, hipMemcpyHostToDevice));
    CHECK_HIP(hipMemcpy(d_B, h_B, size_B, hipMemcpyHostToDevice));

    // Launch kernel
    dim3 threadsPerBlock(16, 16);
    dim3 blocksPerGrid((N + threadsPerBlock.x - 1) / threadsPerBlock.x,
                       (M + threadsPerBlock.y - 1) / threadsPerBlock.y);
    printf("Launching kernel with grid (%d, %d), block (%d, %d)\n",
           blocksPerGrid.x, blocksPerGrid.y, threadsPerBlock.x, threadsPerBlock.y);

    // Run kernel multiple times for sustained high utilization
    for (int iter = 0; iter < 30; iter++) { // Increased iterations
        printf("Iteration %d...\n", iter + 1);
        hipLaunchKernelGGL(matrixMulHeavy, blocksPerGrid, threadsPerBlock, 0, 0,
                           d_A, d_B, d_C, M, N, K);
        CHECK_HIP(hipDeviceSynchronize());
    }

    // Copy result back (optional, for verification)
    CHECK_HIP(hipMemcpy(h_C, d_C, size_C, hipMemcpyDeviceToHost));

    // Basic verification
    printf("Verifying result...\n");
    for (int i = 0; i < 30; i++) { // Reduced checks for speed
        int row = rand() % M;
        int col = rand() % N;
        float sum = 0.0f;
        for (int k = 0; k < K; k++) {
            sum += h_A[row * K + k] * h_B[k * N + col];
        }
        for (int j = 0; j < 500; j++) {
            sum = sum * 1.00001f + sinf(sum) * 0.0001f;
        }
        if (fabs(h_C[row * N + col] - sum) > 1e-2) {
            printf("Verification failed at (%d, %d)!\n", row, col);
            break;
        }
    }
    printf("Test completed successfully\n");

    // Clean up
    CHECK_HIP(hipFree(d_A));
    CHECK_HIP(hipFree(d_B));
    CHECK_HIP(hipFree(d_C));
    free(h_A);
    free(h_B);
    free(h_C);

    return 0;
	#include <hip/hip_runtime.h>
	#include <stdio.h>
	#include <stdlib.h>
	#include <time.h>

	int HIP_DEVICE_ID = 0; // Set to the desired device ID

	#define CHECK_HIP(cmd) { \
	hipError_t err = cmd; \
	if (err != hipSuccess) { \
	fprintf(stderr, "HIP Error: %s at %s:%d\n", hipGetErrorString(err), __FILE__, __LINE__); \
	exit(1); \
	} \
	}



	__global__ void matrixMulHeavy(const float A, const float B, float *C, int M, int N, int K) {
	int row = blockIdx.y * blockDim.y + threadIdx.y;
	int col = blockIdx.x * blockDim.x + threadIdx.x;
	if (row < M && col < N) {
	float sum = 0.0f;
	for (int i = 0; i < K; i++) {
	sum += A[row * K + i] * B[i * N + col];
	}
	// Add heavy compute to maximize GPU utilization
	for (int j = 0; j < 500; j++) { // Increased iterations
	sum = sum * 1.00001f + sinf(sum) * 0.0001f; // Math-heavy operations
	}
	C[row * N + col] = sum;
	}
	}

	int main() {
	// Matrix dimensions: A (M x K), B (K x N), C (M x N)
	int M = 23000;
	int N = 23000;
	int K = 23000;
	size_t size_A = (size_t)M * K * sizeof(float);
	size_t size_B = (size_t)K * N * sizeof(float);
	size_t size_C = (size_t)M * N * sizeof(float);

	printf("Allocating ~%.2f GB VRAM\n", (size_A + size_B + size_C) / (1024.0 * 1024.0 * 1024.0));

	// Set device to Radeon 6700S
	CHECK_HIP(hipSetDevice(HIP_DEVICE_ID));
	hipDeviceProp_t props;
	CHECK_HIP(hipGetDeviceProperties(&props, HIP_DEVICE_ID));
	printf("Running on device: %s\n", props.name);

	// Allocate host memory
	float h_A = (float)malloc(size_A);
	float h_B = (float)malloc(size_B);
	float h_C = (float)malloc(size_C);
	if (!h_A \|\| !h_B \|\| !h_C) {
	fprintf(stderr, "Host memory allocation failed\n");
	exit(1);
	}

	// Initialize matrices
	srand(time(NULL));
	for (size_t i = 0; i < (size_t)M * K; i++) h_A[i] = rand() / (float)RAND_MAX;
	for (size_t i = 0; i < (size_t)K * N; i++) h_B[i] = rand() / (float)RAND_MAX;

	// Allocate device memory
	float d_A, d_B, *d_C;
	CHECK_HIP(hipMalloc(&d_A, size_A));
	CHECK_HIP(hipMalloc(&d_B, size_B));
	CHECK_HIP(hipMalloc(&d_C, size_C));

	// Copy inputs to device
	CHECK_HIP(hipMemcpy(d_A, h_A, size_A, hipMemcpyHostToDevice));
	CHECK_HIP(hipMemcpy(d_B, h_B, size_B, hipMemcpyHostToDevice));

	// Launch kernel
	dim3 threadsPerBlock(16, 16);
	dim3 blocksPerGrid((N + threadsPerBlock.x - 1) / threadsPerBlock.x,
	(M + threadsPerBlock.y - 1) / threadsPerBlock.y);
	printf("Launching kernel with grid (%d, %d), block (%d, %d)\n",
	blocksPerGrid.x, blocksPerGrid.y, threadsPerBlock.x, threadsPerBlock.y);

	// Run kernel multiple times for sustained high utilization
	for (int iter = 0; iter < 30; iter++) { // Increased iterations
	printf("Iteration %d...\n", iter + 1);
	hipLaunchKernelGGL(matrixMulHeavy, blocksPerGrid, threadsPerBlock, 0, 0,
	d_A, d_B, d_C, M, N, K);
	CHECK_HIP(hipDeviceSynchronize());
	}

	// Copy result back (optional, for verification)
	CHECK_HIP(hipMemcpy(h_C, d_C, size_C, hipMemcpyDeviceToHost));

	// Basic verification
	printf("Verifying result...\n");
	for (int i = 0; i < 30; i++) { // Reduced checks for speed
	int row = rand() % M;
	int col = rand() % N;
	float sum = 0.0f;
	for (int k = 0; k < K; k++) {
	sum += h_A[row * K + k] * h_B[k * N + col];
	}
	for (int j = 0; j < 500; j++) {
	sum = sum * 1.00001f + sinf(sum) * 0.0001f;
	}
	if (fabs(h_C[row * N + col] - sum) > 1e-2) {
	printf("Verification failed at (%d, %d)!\n", row, col);
	break;
	}
	}
	printf("Test completed successfully\n");

	// Clean up
	CHECK_HIP(hipFree(d_A));
	CHECK_HIP(hipFree(d_B));
	CHECK_HIP(hipFree(d_C));
	free(h_A);
	free(h_B);
	free(h_C);

	return 0;
No results found