__global__ __shared__ __syncthreads() threadIdx.x blockIdx.x blockDim.x gridDim.x <<<N, 256>>> cudaMalloc cudaMemcpy cudaFree 0x00 warpSize atomicAdd __device__ __constant__ float4 sm_90 0xFF sizeof(float) __global__ __shared__ __syncthreads() threadIdx.x blockIdx.x blockDim.x gridDim.x <<<N, 256>>> cudaMalloc cudaMemcpy cudaFree 0x00 warpSize atomicAdd __device__ __constant__ float4 sm_90 0xFF sizeof(float)
CUDA_KERNELS
GPU_ARCH
THREAD_BLOCKS
HPC_SYSTEMS
AI_INFRA
PARALLEL_EXECUTION
/// UNIVERSITY OF HOUSTON
Parallel Computing Society The Future of Computing Is Parallel.
__global__ void vectorAdd( float * a, float * b, float * c, int n) // int idx = blockIdx.x * blockDim.x + threadIdx.x; // cudaMalloc (( void **)&d_data, size * sizeof( float )); // cudaMemcpy (d_data, h_data, size, cudaMemcpyHostToDevice ); // __shared__ float tile[TILE_DIM][TILE_DIM]; // dim3 threadsPerBlock(16, 16); // matMulKernel<<<numBlocks, threadsPerBlock>>>(A, B, C, N); // __syncthreads (); // if (idx < n) c[idx] = a[idx] + b[idx]; // atomicAdd (&output[idx], value); // cudaDeviceSynchronize (); // cudaStreamCreate (&stream); // #pragma unroll // nvcc -arch=sm_90 -O3 kernel.cu -o kernel // cublasSgemm (handle, CUBLAS_OP_N, CUBLAS_OP_N, m, n, k, ...); // warpReduceSum (val); // butterfly shuffle // cudaFree (d_data); // __global__ void vectorAdd( float * a, float * b, float * c, int n) // int idx = blockIdx.x * blockDim.x + threadIdx.x; // cudaMalloc (( void **)&d_data, size * sizeof( float )); // cudaMemcpy (d_data, h_data, size, cudaMemcpyHostToDevice ); // __shared__ float tile[TILE_DIM][TILE_DIM]; // dim3 threadsPerBlock(16, 16); // matMulKernel<<<numBlocks, threadsPerBlock>>>(A, B, C, N); // __syncthreads (); // if (idx < n) c[idx] = a[idx] + b[idx]; // atomicAdd (&output[idx], value); // cudaDeviceSynchronize (); // cudaStreamCreate (&stream); // #pragma unroll // nvcc -arch=sm_90 -O3 kernel.cu -o kernel // cublasSgemm (handle, CUBLAS_OP_N, CUBLAS_OP_N, m, n, k, ...); // warpReduceSum (val); // butterfly shuffle // cudaFree (d_data); //
__global__ __shared__ __syncthreads() threadIdx.x blockIdx.x blockDim.x gridDim.x <<<N, 256>>> cudaMalloc cudaMemcpy cudaFree 0x00 warpSize atomicAdd __device__ __constant__ float4 sm_90 0xFF sizeof(float) __global__ __shared__ __syncthreads() threadIdx.x blockIdx.x blockDim.x gridDim.x <<<N, 256>>> cudaMalloc cudaMemcpy cudaFree 0x00 warpSize atomicAdd __device__ __constant__ float4 sm_90 0xFF sizeof(float)
__global__ __shared__ __syncthreads() threadIdx.x blockIdx.x blockDim.x gridDim.x <<<N, 256>>> cudaMalloc cudaMemcpy cudaFree 0x00 warpSize atomicAdd __device__ __constant__ float4 sm_90 0xFF sizeof(float) __global__ __shared__ __syncthreads() threadIdx.x blockIdx.x blockDim.x gridDim.x <<<N, 256>>> cudaMalloc cudaMemcpy cudaFree 0x00 warpSize atomicAdd __device__ __constant__ float4 sm_90 0xFF sizeof(float)
SEC.01 // MISSION
Our Mission UH PCS closes the gap between traditional CS/ECE education and the real-world demand for GPU computing, parallel systems, and performance-focused engineering.
pcs://mission/01_learn
01_LEARN
Hands-on workshops in CUDA, GPU architecture, and systems thinking.
pcs://mission/02_build
02_BUILD
Real projects that turn concepts into portfolio-ready technical work.
pcs://mission/03_connect
03_CONNECT
Industry talks, recruiter access, alumni support, and mentorship.
__global__ void vectorAdd( float * a, float * b, float * c, int n) // int idx = blockIdx.x * blockDim.x + threadIdx.x; // cudaMalloc (( void **)&d_data, size * sizeof( float )); // cudaMemcpy (d_data, h_data, size, cudaMemcpyHostToDevice ); // __shared__ float tile[TILE_DIM][TILE_DIM]; // dim3 threadsPerBlock(16, 16); // matMulKernel<<<numBlocks, threadsPerBlock>>>(A, B, C, N); // __syncthreads (); // if (idx < n) c[idx] = a[idx] + b[idx]; // atomicAdd (&output[idx], value); // cudaDeviceSynchronize (); // cudaStreamCreate (&stream); // #pragma unroll // nvcc -arch=sm_90 -O3 kernel.cu -o kernel // cublasSgemm (handle, CUBLAS_OP_N, CUBLAS_OP_N, m, n, k, ...); // warpReduceSum (val); // butterfly shuffle // cudaFree (d_data); // __global__ void vectorAdd( float * a, float * b, float * c, int n) // int idx = blockIdx.x * blockDim.x + threadIdx.x; // cudaMalloc (( void **)&d_data, size * sizeof( float )); // cudaMemcpy (d_data, h_data, size, cudaMemcpyHostToDevice ); // __shared__ float tile[TILE_DIM][TILE_DIM]; // dim3 threadsPerBlock(16, 16); // matMulKernel<<<numBlocks, threadsPerBlock>>>(A, B, C, N); // __syncthreads (); // if (idx < n) c[idx] = a[idx] + b[idx]; // atomicAdd (&output[idx], value); // cudaDeviceSynchronize (); // cudaStreamCreate (&stream); // #pragma unroll // nvcc -arch=sm_90 -O3 kernel.cu -o kernel // cublasSgemm (handle, CUBLAS_OP_N, CUBLAS_OP_N, m, n, k, ...); // warpReduceSum (val); // butterfly shuffle // cudaFree (d_data); //
__global__ __shared__ __syncthreads() threadIdx.x blockIdx.x blockDim.x gridDim.x <<<N, 256>>> cudaMalloc cudaMemcpy cudaFree 0x00 warpSize atomicAdd __device__ __constant__ float4 sm_90 0xFF sizeof(float) __global__ __shared__ __syncthreads() threadIdx.x blockIdx.x blockDim.x gridDim.x <<<N, 256>>> cudaMalloc cudaMemcpy cudaFree 0x00 warpSize atomicAdd __device__ __constant__ float4 sm_90 0xFF sizeof(float)
__global__ __shared__ __syncthreads() threadIdx.x blockIdx.x blockDim.x gridDim.x <<<N, 256>>> cudaMalloc cudaMemcpy cudaFree 0x00 warpSize atomicAdd __device__ __constant__ float4 sm_90 0xFF sizeof(float) __global__ __shared__ __syncthreads() threadIdx.x blockIdx.x blockDim.x gridDim.x <<<N, 256>>> cudaMalloc cudaMemcpy cudaFree 0x00 warpSize atomicAdd __device__ __constant__ float4 sm_90 0xFF sizeof(float)
SEC.02 // ACTIVE_PROTOCOLS
What We Offer From first CUDA kernel to portfolio-ready systems.
__global__ void vectorAdd( float * a, float * b, float * c, int n) // int idx = blockIdx.x * blockDim.x + threadIdx.x; // cudaMalloc (( void **)&d_data, size * sizeof( float )); // cudaMemcpy (d_data, h_data, size, cudaMemcpyHostToDevice ); // __shared__ float tile[TILE_DIM][TILE_DIM]; // dim3 threadsPerBlock(16, 16); // matMulKernel<<<numBlocks, threadsPerBlock>>>(A, B, C, N); // __syncthreads (); // if (idx < n) c[idx] = a[idx] + b[idx]; // atomicAdd (&output[idx], value); // cudaDeviceSynchronize (); // cudaStreamCreate (&stream); // #pragma unroll // nvcc -arch=sm_90 -O3 kernel.cu -o kernel // cublasSgemm (handle, CUBLAS_OP_N, CUBLAS_OP_N, m, n, k, ...); // warpReduceSum (val); // butterfly shuffle // cudaFree (d_data); // __global__ void vectorAdd( float * a, float * b, float * c, int n) // int idx = blockIdx.x * blockDim.x + threadIdx.x; // cudaMalloc (( void **)&d_data, size * sizeof( float )); // cudaMemcpy (d_data, h_data, size, cudaMemcpyHostToDevice ); // __shared__ float tile[TILE_DIM][TILE_DIM]; // dim3 threadsPerBlock(16, 16); // matMulKernel<<<numBlocks, threadsPerBlock>>>(A, B, C, N); // __syncthreads (); // if (idx < n) c[idx] = a[idx] + b[idx]; // atomicAdd (&output[idx], value); // cudaDeviceSynchronize (); // cudaStreamCreate (&stream); // #pragma unroll // nvcc -arch=sm_90 -O3 kernel.cu -o kernel // cublasSgemm (handle, CUBLAS_OP_N, CUBLAS_OP_N, m, n, k, ...); // warpReduceSum (val); // butterfly shuffle // cudaFree (d_data); //
__global__ __shared__ __syncthreads() threadIdx.x blockIdx.x blockDim.x gridDim.x <<<N, 256>>> cudaMalloc cudaMemcpy cudaFree 0x00 warpSize atomicAdd __device__ __constant__ float4 sm_90 0xFF sizeof(float) __global__ __shared__ __syncthreads() threadIdx.x blockIdx.x blockDim.x gridDim.x <<<N, 256>>> cudaMalloc cudaMemcpy cudaFree 0x00 warpSize atomicAdd __device__ __constant__ float4 sm_90 0xFF sizeof(float)
__global__ __shared__ __syncthreads() threadIdx.x blockIdx.x blockDim.x gridDim.x <<<N, 256>>> cudaMalloc cudaMemcpy cudaFree 0x00 warpSize atomicAdd __device__ __constant__ float4 sm_90 0xFF sizeof(float) __global__ __shared__ __syncthreads() threadIdx.x blockIdx.x blockDim.x gridDim.x <<<N, 256>>> cudaMalloc cudaMemcpy cudaFree 0x00 warpSize atomicAdd __device__ __constant__ float4 sm_90 0xFF sizeof(float)
SEC.03 // SOURCE_CONTROL
Open Repositories. Real Systems. PCS projects are built to be studied, forked, improved, and shown. Our GitHub is the technical portfolio of the organization.
uh-pcs@github:~$ ls repositories/ --featured
8-Bit CPU Emulator A cycle-accurate emulator teaching fetch, decode, execute, and registers.
C++ Assembly Concepts
$ git clone uh-pcs/future-of-computing
__global__ void vectorAdd( float * a, float * b, float * c, int n) // int idx = blockIdx.x * blockDim.x + threadIdx.x; // cudaMalloc (( void **)&d_data, size * sizeof( float )); // cudaMemcpy (d_data, h_data, size, cudaMemcpyHostToDevice ); // __shared__ float tile[TILE_DIM][TILE_DIM]; // dim3 threadsPerBlock(16, 16); // matMulKernel<<<numBlocks, threadsPerBlock>>>(A, B, C, N); // __syncthreads (); // if (idx < n) c[idx] = a[idx] + b[idx]; // atomicAdd (&output[idx], value); // cudaDeviceSynchronize (); // cudaStreamCreate (&stream); // #pragma unroll // nvcc -arch=sm_90 -O3 kernel.cu -o kernel // cublasSgemm (handle, CUBLAS_OP_N, CUBLAS_OP_N, m, n, k, ...); // warpReduceSum (val); // butterfly shuffle // cudaFree (d_data); // __global__ void vectorAdd( float * a, float * b, float * c, int n) // int idx = blockIdx.x * blockDim.x + threadIdx.x; // cudaMalloc (( void **)&d_data, size * sizeof( float )); // cudaMemcpy (d_data, h_data, size, cudaMemcpyHostToDevice ); // __shared__ float tile[TILE_DIM][TILE_DIM]; // dim3 threadsPerBlock(16, 16); // matMulKernel<<<numBlocks, threadsPerBlock>>>(A, B, C, N); // __syncthreads (); // if (idx < n) c[idx] = a[idx] + b[idx]; // atomicAdd (&output[idx], value); // cudaDeviceSynchronize (); // cudaStreamCreate (&stream); // #pragma unroll // nvcc -arch=sm_90 -O3 kernel.cu -o kernel // cublasSgemm (handle, CUBLAS_OP_N, CUBLAS_OP_N, m, n, k, ...); // warpReduceSum (val); // butterfly shuffle // cudaFree (d_data); //
__global__ __shared__ __syncthreads() threadIdx.x blockIdx.x blockDim.x gridDim.x <<<N, 256>>> cudaMalloc cudaMemcpy cudaFree 0x00 warpSize atomicAdd __device__ __constant__ float4 sm_90 0xFF sizeof(float) __global__ __shared__ __syncthreads() threadIdx.x blockIdx.x blockDim.x gridDim.x <<<N, 256>>> cudaMalloc cudaMemcpy cudaFree 0x00 warpSize atomicAdd __device__ __constant__ float4 sm_90 0xFF sizeof(float)
__global__ __shared__ __syncthreads() threadIdx.x blockIdx.x blockDim.x gridDim.x <<<N, 256>>> cudaMalloc cudaMemcpy cudaFree 0x00 warpSize atomicAdd __device__ __constant__ float4 sm_90 0xFF sizeof(float) __global__ __shared__ __syncthreads() threadIdx.x blockIdx.x blockDim.x gridDim.x <<<N, 256>>> cudaMalloc cudaMemcpy cudaFree 0x00 warpSize atomicAdd __device__ __constant__ float4 sm_90 0xFF sizeof(float)
pcs://join/access_control
SEC.04 // ACCESS_CONTROL
Start before you feel ready. You do not need CUDA experience to join. PCS is built for students who are curious, ambitious, and ready to learn by building.