contribute some solutions to easy problems that are not yet in this repository.

[wenj-yan]

Value Clipping

Problem

Given an array input of N floats and bounds lo, hi, clamp each element to [lo, hi]:
output[i] = min(max(input[i], lo), hi)

Baseline

__global__ void clip_kernel(const float* input, float* output, float lo, float hi, int N) {
    int idx = blockDim.x * blockIdx.x + threadIdx.x;
    if (idx >= N) return;

    float out = input[idx] > lo ? input[idx] : lo;
    output[idx] = out < hi ? out : hi;
}
One thread per element. Simple but inefficient — memory‑bound workload with 32‑bit transactions.

Optimized (Vectorized)

__global__ void clip_kernel(const float* input, float* output, float lo, float hi, int N) {
    int idx = blockDim.x * blockIdx.x + threadIdx.x;
    if (idx > N / 4) return;

    if (idx == N / 4) {
        for (int i = 0; i < N % 4; i++) {
            float tmp = input[idx * 4 + i];
            float out = tmp > lo ? tmp : lo;
            out = out < hi ? out : hi;
            output[idx * 4 + i] = out;
        }
    } else if (idx < N / 4) {
        const float4 tmp = reinterpret_cast<const float4*>(input)[idx];
        float4 out;
        out.x = tmp.x > lo ? tmp.x : lo;
        out.x = out.x < hi ? out.x : hi;
        out.y = tmp.y > lo ? tmp.y : lo;
        out.y = out.y < hi ? out.y : hi;
        out.z = tmp.z > lo ? tmp.z : lo;
        out.z = out.z < hi ? out.z : hi;
        out.w = tmp.w > lo ? tmp.w : lo;
        out.w = out.w < hi ? out.w : hi;
        reinterpret_cast<float4*>(output)[idx] = out;
    }
}