一只大笨熊

SoftMax Softmax 的 CPU 和 CUDA 写法均是高频考察。面试时有可能会让任选一种写法进行书写，此时自己可以先写 CPU(C++、Python) 版本，然后再写 CUDA 版本。 Softmax公式如下：softmax(xi)=exi∑jexjsoftmax(x_i) = \frac{e^{x_i}}{\sum_j e^{x_j}} softmax(xi​)=∑j​exj​exi​​ 一般为了避免溢出，需要减去最大值，所以通常采用下面这个公式：softmax(xi)=exi−max(x)∑jexj−max(x)softmax(x_i) = \frac{e^{x_i - max(x)}}{\sum_j e^{x_j - max(x)}} softmax(xi​)=∑j​exj​−max(x)exi​−max(x)​ 1. CPU(C++、Python) 版本 1234567891011void softmax(float* input, float* output, int N){ float max_value =...

elementwise elementwise 是最简单的一类算子，其指的是对数据进行逐元素操作，例如将两个等长的数组对应元素相加（add）。另外在深度学习中，激活函数会对输入数据的每个元素求对应激活值，故激活函数也算在 elementwise 范围内。 add 1234567891011121314151617181920212223242526272829303132333435363738394041// 1. 向上取整#define CEIL(a, b) ((a + b - 1) / (b))// 2. FLOAT4，用于向量化访存，以下两种都可以// c写法#define FLOAT4(value) *(float4*)(&(value))// c++写法#define FLOAT4(value) (reinterpret_cast<float4*>(&(value))[0])//naive版int block_size = 1024;int grid_size = CEIL(N,...

Reduce 算子是指通过对数组中的每个元素进行操作，得到一个输出值的过程。常见的操作包括求和（sum）、取最大值（max）、取最小值（min）等。在 CUDA 中，优化 Reduce 算子可以显著提高计算效率。 1. naive实现 1234567//累加__global__ void reduce1(float* d_in, float* d_out, int N) { int idx = blockIdx.x * blockDim.x + threadIdx.x; if (idx < N) { atomicAdd(d_out, d_in[idx]); }} 2. share mem + 折半规约 12345678910111213141516171819202122232425262728__global__ void reduce2(float* d_in, float* d_out, int N) { __shared__ float sdata[BLOCK_SIZE]; ...

1. cpu: 矩阵乘法 1234567891011121314151617181920212223242526272829// 二维矩阵void matrixMultiply(const float** A, const float** B, float** C, int m, int p, int n) { // A is m x p, B is p x n, C is m x n for (int i = 0; i < m; ++i) { for (int j = 0; j < n; ++j) { float sum = 0.0; for (int k = 0; k < p; ++k) { sum += A[i][k] * B[k][j]; } C[i][j] = sum; } }}// 二维矩阵展开成一维void...

1.cpu: 计算每行的softmax 12345678910111213141516void softmax_row(float* input, float* output, int M, int N) { for (int row = 0; row < M; row++) { // 第row行 float* input_tmp = input + row * N; float* output_tmp = output + row * N; float max_val = *(std::max_element(input_tmp, input_tmp + N)); // 计算输入数组的最大值 float sum = 0; for (int i = 0; i < N; i++) { output_tmp[i] = std::exp(input_tmp[i] - max_val); //...

naive版本 1234567__global__ void transpose_v0(float* input, float* output, int M, int N){ int row = blockIdx.y * blockDim.y + threadIdx.y; int col = blockIdx.x * blockDim.x + threadIdx.x; if(row < M && col < N){ output[col * M + row] = input[row * N + col]; }} 优化版本1：shared memory 思路： 先将数据从global memory拷贝到shared memory中 通过shared memory进行转置 通过shared memory将数据拷贝到global memory中 1234567891011121314151617template <int TILE_SIZE>__global__...

使用C++实现一个读写锁 实现一个shared_ptr 123456789101112131415161718192021222324252627282930313233343536373839404142434445464748495051525354555657585960616263646566676869707172737475767778798081828384858687888990919293949596979899100101102103104105#pragma once#include <utility>#include <cstddef>template<typename T>class SimpleSharedPtr {private: struct ControlBlock { T* ptr; size_t count; explicit ControlBlock(T* p) : ptr(p), count(1)...

第1章 计算机网络体系结构 第2章 物理层 第3章 数据链路层 第4章 网络层 NAT 的作用是什么？ NAT（Network Address Translation，网络地址转换） 主要用于在不同网络之间转换 IP 地址。它允许将私有 IP 地址（如在局域网中使用的 IP 地址）映射为公有 IP 地址（在互联网中使用的 IP 地址）或者反向映射，从而实现局域网内的多个设备通过单一公有 IP 地址访问互联网。 NAT 不光可以缓解 IPv4...

1. 实现LRUCache 123456789101112131415161718192021222324252627282930313233343536373839404142434445464748495051525354555657585960616263646566676869707172737475767778798081828384858687888990919293949596979899100101102103104105106107108109110111112113114115116117118119120121122123124125126127128129130131132133134135136// 方法一：不使用STL，自己设计数据结构#include<unordered_map>#include<iostream>using namespace std;struct DLinkedNode { int key, value; DLinkedNode* prev; DLinkedNode*...

...